GPU計算の三銃士シリーズ第2回

「行列の掛け算なんて、二重ループで書けばいいでしょ？」

CUDA を勉強し始めた頃の自分は、本気でそう思っていた。実際、for を二つ重ねれば動く。動くんだが、ベンチマークを取って真顔になった。自作版とcuBLASで、軽く100倍以上の差が出た(;ﾟдﾟ)ﾎﾟｶｰﾝ

orz...と崩れ落ちた経験から学んだのは、行列演算の世界では「動く」と「使える」の間にグランドキャニオン級の谷があるということだ。

今回は、GPU上で行列演算を異次元のスピードで実行するライブラリ cuBLAS を解説する。50年の歴史を持つBLAS仕様の話から、最新Blackwell世代のTensor Core活用まで、一気通貫で見ていこう。

この記事の対象読者

PyTorchやTensorFlowを使っているが、その下で何が動いているか知りたい人
自作CUDAコードと既存ライブラリのどちらを使うべきか迷っている人
ディープラーニングが速い理由を「行列演算レベル」で理解したい人
「BLAS」「GEMM」「Tensor Core」の関係を整理したい人

この記事で得られること

BLASという「50年続く標準仕様」と cuBLAS の関係が分かる
GEMM がディープラーニングの中核を担っている理由が分かる
cuBLAS / cuBLASLt / cuBLASXt / cuBLASMp / cuBLASDx の使い分けが分かる
自作 vs ライブラリの判断軸を、性能・信頼性・保守性で説明できる
パフォーマンス最適化の基本(8の倍数ルール、混合精度、バッチGEMM)を実践できる

この記事で扱わないこと

BLAS のFORTRAN実装の内部詳細(歴史的経緯のみ触れる)
cuDNN や cuFFT などの他CUDAライブラリ(次回以降のシリーズで扱う)
自作 GEMM カーネルの最適化テクニック(別記事のスコープ)
ROCm / hipBLAS など AMD GPU 向け実装

TL;DR

BLAS = 線形代数演算の世界標準仕様。1979年から続く「ISO規格のようなもの」
cuBLAS = NVIDIA が GPU 向けに最適化したBLAS実装
自作の二重ループより数十倍〜数百倍速い
速さの正体は Tensor Core と アーキテクチャ別チューニング
PyTorchやTensorFlow を使っていれば、知らぬ間に cuBLAS が走っている

1. 統一比喩: cuBLAS は「NVIDIA 印の自動車工場」だ

この記事全体を通して、cuBLAS を 自動車工場 に喩えて説明する。記事を読み終わる頃には、行列演算とトヨタの生産方式の親近感がきっと深まっているはずだ。

cuBLAS の概念	自動車工場での対応物
BLAS(仕様)	自動車部品の世界標準規格(ISO)
cuBLAS(実装)	NVIDIA 印の最新自動化工場
GEMM	工場で最も重要な主要生産ライン
Tensor Core	高速ロボットアーム
バッチGEMM	同型ラインを並列稼働する大量生産モード
cuBLASLt	カスタムオーダー対応の高級ライン
自作行列積	個人ガレージでの手作業組立

ガレージで「車くらい自分で作れるよ」と言っている人と、最新自動化工場の生産ラインを比べたら、結果は見えている。それと同じ話だ。

2. BLAS とは何か — 50年続く「部品規格」

2.1 1970年代の英知

cuBLAS の話に入る前に、まず BLAS という土台を押さえておく。

BLAS は Basic Linear Algebra Subprograms の略で、線形代数の基本操作を標準化した仕様だ。

"Basic Linear Algebra Subprograms (BLAS) is a specification that prescribes a set of low-level routines for performing common linear algebra operations such as vector addition, scalar multiplication, dot products, linear combinations, and matrix multiplication."

「BLAS(Basic Linear Algebra Subprograms)は、ベクトル加算、スカラー倍、内積、線形結合、行列乗算などの一般的な線形代数演算を行うための低レベルルーチン群を規定した仕様である。」

— Wikipedia "Basic Linear Algebra Subprograms" (https://en.wikipedia.org/wiki/Basic_Linear_Algebra_Subprograms)

工場の比喩で言えば、BLAS は「ボルトのサイズはM8で、ねじピッチは1.25mm」と決めた 世界標準のような存在 に相当する。この規格があるから、世界中のメーカーが互換性のある部品を作れる。

"A specification for these kernel operations using scalars and vectors, the level-1 Basic Linear Algebra Subroutines (BLAS), was published in 1979."

「スカラーとベクトルを使ったこれらのカーネル操作の仕様、Level-1 BLAS(Basic Linear Algebra Subroutines)は1979年に公開された。」

— Wikipedia "Basic Linear Algebra Subprograms" (https://en.wikipedia.org/wiki/Basic_Linear_Algebra_Subprograms)

2.2 3つのレベル — 部品から完成車まで

BLAS は操作の種類によって3つのレベルに分かれている。これは工場で言えば「ボルト製造」「サブアセンブリ」「車体組立」という工程の階層に近い。

レベル	操作の種類	例	計算量	工場の比喩
Level 1	ベクトル–ベクトル	内積、加算	O(n)	ボルト単品の製造
Level 2	行列–ベクトル	行列とベクトルの積	O(n²)	サブアセンブリ(部品の組み合わせ)
Level 3	行列–行列	行列積(GEMM)	O(n³)	車体の最終組立

"The Level 1 BLAS perform scalar, vector and vector-vector operations, the Level 2 BLAS perform matrix-vector operations, and the Level 3 BLAS perform matrix-matrix operations."

「Level 1 BLAS はスカラー・ベクトル・ベクトル-ベクトル演算を、Level 2 BLAS は行列-ベクトル演算を、Level 3 BLAS は行列-行列演算を行う。」

— Netlib BLAS (https://www.netlib.org/blas/)

計算量が指数的に増えるため、Level 3 をいかに速く回せるかが工場の総生産力を決める。後述する GEMM はまさにこの Level 3 の中核だ。

2.3 仕様と実装の分離 — マルチベンダー時代の知恵

BLAS の最大の発明は、仕様と実装を分離した こと。同じインターフェースを使いながら、ハードウェアごとに最適化された実装に差し替えられる。

これは「ISO規格のボルトなら、トヨタ製でもデンソー製でも交換可能」というのと同じ発想だ。

実装	ベンダー	対象
MKL	Intel	Intel CPU
AOCL	AMD	AMD CPU
OpenBLAS	OSS	汎用CPU
cuBLAS	NVIDIA	NVIDIA GPU
rocBLAS	AMD	AMD GPU

つまり、コードは1行も変えずに、リンクするライブラリを切り替えるだけで速くなる可能性がある。これが標準化の威力だ。

3. cuBLAS の正体 — NVIDIA 印の最新自動化工場

3.1 GPU 向け実装の本命

cuBLAS は、NVIDIA が自社 GPU 向けに極限まで最適化した BLAS 実装 だ。

"The cuBLAS library is an implementation of BLAS (Basic Linear Algebra Subprograms) on top of the NVIDIA® CUDA™ runtime."

「cuBLAS ライブラリは、NVIDIA CUDA ランタイム上に構築された BLAS の実装である。」

— cuBLAS Documentation (https://docs.nvidia.com/cuda/cublas/)

CUDA Toolkit に同梱されているので、CUDA をインストールすれば追加作業なしで使える。

3.2 Tensor Core — 高速ロボットアームの登場

Volta 世代(2017年)以降の NVIDIA GPU には、Tensor Core という行列演算専用のハードウェアが搭載されている。比喩で言えば、汎用ロボットアーム(CUDA Core)の隣に、特定の組立作業だけを猛烈な速度でこなす特殊ロボット が配備されたイメージだ。

"The cuBLAS library is highly optimized for performance on NVIDIA GPUs, and leverages tensor cores for acceleration of low- and mixed-precision matrix multiplication."

「cuBLAS ライブラリは NVIDIA GPU 上での性能に高度に最適化されており、低精度および混合精度の行列乗算を加速するために Tensor Core を活用する。」

— NVIDIA cuBLAS (https://developer.nvidia.com/cublas)

最新の RTX 5090 や RTX Pro 6000 では、この Tensor Core がさらに進化し、FP4 / FP8 などの超低精度演算にも対応している。

4. GEMM — 工場で最も忙しい生産ライン

4.1 GEMM とは

cuBLAS で最も重要な操作が GEMM 、すなわち General Matrix Multiply だ。式で書くと:

C = \alpha A B + \beta C

ここで A, B, C は行列、α, β はスカラー。単純な行列積 C = AB は α=1, β=0 の特殊ケースになる。

4.2 なぜ GEMM が「主要ライン」なのか

驚くべきことに、ディープラーニングの計算の大部分は GEMM に帰着する。

演算	GEMM への帰着
全結合層(Dense)	入力ベクトル × 重み行列
畳み込み層(Conv)	im2col で行列展開 → GEMM
アテンション	Q · Kᵀ、softmax後 · V のすべてが行列積
バッチ処理	サンプルをまとめて1つの大きな行列に

つまり、GEMM が速ければディープラーニング全体が速くなる。工場で言えば、エンジン組立ラインのスループットが工場全体の生産力を決めるのと同じ構造だ。

"Two CUDA libraries that use Tensor Cores are cuBLAS and cuDNN. cuBLAS uses Tensor Cores to speed up GEMM computations."

「Tensor Core を使用する2つの CUDA ライブラリが cuBLAS と cuDNN だ。cuBLAS は Tensor Core を使って GEMM 計算を高速化する。」

— NVIDIA Developer Blog (https://developer.nvidia.com/blog/programming-tensor-cores-cuda-9/)

5. cuBLAS ファミリー — ライン構成のバリエーション

cuBLAS には用途に応じた複数のバリエーションがある。同じ工場でも「軽自動車ライン」「高級セダンライン」「商用車ライン」と分かれているようなものだ。

バリエーション	役割	工場の比喩
cuBLAS	標準版。Level 1–3 全対応	標準生産ライン
cuBLASLt	柔軟な GEMM API、Tensor Core 詳細制御、エピローグ融合	カスタムオーダー対応の高級ライン
cuBLASXt	単一ホストの複数 GPU 分散	同一工場敷地内の複数ライン同時稼働
cuBLASMp(Preview)	マルチノード・マルチGPU(HPC向け)	複数工場を結んだ生産ネットワーク
cuBLASDx(Preview)	CUDA カーネル内から直接呼べる	ライン内部に小型自動装置を組み込み

5.1 cuBLASLt の存在意義

「Lt」は Lightweight の略。ただし「機能が軽い」のではなく、より柔軟で表現力が高い という意味だ。

"cuBLASLt Host APIs are multi-stage GEMM APIs that are highly expressive, allowing applications to leverage the latest NVIDIA architecture features for the best performance with support for fusions and performance tuning options."

「cuBLASLt ホストAPIは、マルチステージの GEMM API であり、表現力が高く、最新の NVIDIA アーキテクチャ機能を活用しつつ、フュージョンやパフォーマンスチューニングオプションをサポートして最高性能を達成できる。」

— NVIDIA cuBLAS (https://developer.nvidia.com/cublas)

主な強み:

Tensor Core の詳細制御(アルゴリズム選択など)
エピローグ融合(GEMM の後に Bias 加算や ReLU を一気に処理)
最新アーキテクチャの新機能をいち早く利用可能

5.2 cuBLASDx — カーネル内に組み込む小型ライン

"cuBLASDx (Preview) is a device side API extension to cuBLAS for performing BLAS calculations inside your CUDA kernel. Fusing numerical operations decreases the latency and improves the performance of your application."

「cuBLASDx(プレビュー)は、CUDA カーネル内で BLAS 計算を行うための cuBLAS のデバイスサイド API 拡張である。数値演算のフュージョンによりレイテンシが下がり、アプリケーションの性能が向上する。」

— NVIDIA cuBLAS (https://developer.nvidia.com/cublas)

通常の cuBLAS はホスト側から呼ぶ(ライン全体を外から発注)が、cuBLASDx は カーネル内部に直接組み込める ので、複数の演算をまとめてレイテンシを削れる。

6. 自作 vs cuBLAS — ガレージ製造の限界

6.1 「書けるけど、書くべきではない」

Stack Overflow で「行列積くらい自分で書けるのでは?」という質問がよくある。答えは明確だ — 書けるが、書くべきではない。

"It will be optimised. NVIDIA supported libraries such as cuBLAS are likely to be optimised for all current GPU generations, and later releases will be optimised for later generations. While most BLAS operations may seem fairly simple to implement, to get peak performance you have to optimise for hardware. A simple implementation of SGEMM, for example, may be many times slower than an optimised version."

「最適化されている。cuBLAS のような NVIDIA サポートのライブラリは、現行すべての GPU 世代に最適化されており、将来のリリースは将来世代に最適化される。BLAS 操作は実装が簡単に見えるかもしれないが、ピーク性能を出すにはハードウェアへの最適化が必要だ。例えば SGEMM の単純実装は、最適化版より何倍も遅くなりうる。」

— Stack Overflow (https://stackoverflow.com/questions/25836003/normal-cuda-vs-cublas)

これを工場の比喩で言うとこうなる。

観点	ガレージ手作業(自作)	最新自動化工場(cuBLAS)
1台あたりの時間	数ヶ月	数十秒
品質保証	職人の腕次第	ISO認証・統計的品質管理
新型対応	モデルごとに一から学習	自動アップデート
故障時の対応	自分でデバッグ	公式サポート

6.2 cuBLAS を使うべき理由 — 4つの根拠

最適化の深さ: NVIDIA のエンジニアは自社 GPU の内部仕様を完全に把握している。メモリアクセスパターン、キャッシュ挙動、Tensor Core 活用、ドキュメントに載らないレベルの最適化が組み込まれている
世代間の自動対応: 新アーキテクチャが出るたびに自動でチューニング済みになる。自作だと毎回チューニングが必要
信頼性: 数百万のユーザーが使い込んでいるコードベース。自作のバグを踏むリスクの方がずっと高い
開発時間の節約: 車輪の再発明をやめると、本来解決すべき問題に集中できる

"They tend to work. There's probably less chance you'll run up against a bug in a library than you'll create a bug in your own implementation."

「ちゃんと動く。ライブラリのバグを踏む確率より、自作実装でバグを作る確率の方が高い。」

— Stack Overflow (https://stackoverflow.com/questions/25836003/normal-cuda-vs-cublas)

7. 使い方 — ライン稼働の手順

7.1 初期化と終了(C++ ネイティブ)

クリックでサンプルコードを展開

#include <cublas_v2.h>
#include <cuda_runtime.h>

int main() {
    cublasHandle_t handle;
    cublasStatus_t status = cublasCreate(&handle);
    if (status != CUBLAS_STATUS_SUCCESS) {
        // 工場の管理棟を立ち上げ失敗
        return -1;
    }

    // ... ここで GEMM などを呼ぶ ...

    cublasDestroy(handle);
    return 0;
}

7.2 SGEMM — 単精度行列積の例

クリックで SGEMM 呼び出し例を展開

// C = alpha * A * B + beta * C
float alpha = 1.0f;
float beta  = 0.0f;

cublasSgemm(handle,
    CUBLAS_OP_N, CUBLAS_OP_N,   // 転置なし
    m, n, k,                    // 行列サイズ
    &alpha,
    d_A, lda,                   // 行列A (デバイスメモリ)
    d_B, ldb,                   // 行列B
    &beta,
    d_C, ldc);                  // 結果C

注意: cuBLAS は 列優先 すなわち Column-major で動く。C/C++ の通常の行優先(Row-major)とは逆なので、lda/ldb/ldc の解釈に気をつける。

7.3 PyTorch からの利用 — 知らずに使っている

PyTorch ユーザーにとって朗報。何もしなくても cuBLAS は呼ばれている。

import torch

# GPU 上で行列積を計算 → 内部で cuBLAS の GEMM が走る
A = torch.randn(1024, 1024, device='cuda')
B = torch.randn(1024, 1024, device='cuda')
C = torch.mm(A, B)  # ← cuBLAS Sgemm が呼ばれる

torch.matmul、nn.Linear、F.conv2d の内部、Transformer のアテンション層 — どれも cuBLAS の世話になっている。

8. パフォーマンス最適化 — 工場の生産効率を上げる3つの鉄則

8.1 鉄則1: 行列サイズを 8 の倍数にする

Tensor Core を効率的に使うには、行列の次元が FP16 なら8の倍数、INT8 なら16の倍数 である必要がある。

"We recommend ensuring all such parameters are multiples of 8 when training with FP16 and multiples of 16 when training with INT8."

「FP16 で訓練する場合はこれらのパラメータをすべて8の倍数に、INT8 で訓練する場合は16の倍数にすることを推奨する。」

— NVIDIA Developer Blog (https://developer.nvidia.com/blog/optimizing-gpu-performance-tensor-cores/)

例: バッチサイズ 30 → 32 にするだけで Tensor Core が活性化することがある。ラインの規格に合わせて部品を揃える、という工場の基本動作と同じだ。

8.2 鉄則2: 混合精度で帯域を稼ぐ

FP16(半精度)を使うと:

項目	効果
Tensor Core	フル活用
メモリ使用量	FP32 比で半減
メモリ帯域	実効的に倍増
精度	重要な箇所のみ FP32 で計算(自動)

# PyTorch の自動混合精度
from torch.cuda.amp import autocast, GradScaler

scaler = GradScaler()

with autocast():
    output = model(input)
    loss = criterion(output, target)

scaler.scale(loss).backward()
scaler.step(optimizer)
scaler.update()

8.3 鉄則3: 小さい行列はバッチ GEMM でまとめる

小さな行列を大量に処理する場合、1回ずつ呼ぶと API オーバーヘッドが効いてくる。バッチ GEMM で一気に流す方が効率的だ。

工場で言えば「1台ずつ生産ラインを起動して停止」を繰り返すより、「100台分まとめて流す」方が早い、という話。

9. 決定論性と再現性 — 工場の品質管理

科学計算や検証で重要な「同じ入力 → 同じ結果」を保証できるか、という話。

"By design, all cuBLAS API routines from a given toolkit version, generate the same bit-wise results at every run when executed on GPUs with the same architecture and the same number of SMs."

「設計上、特定のToolkitバージョンの cuBLAS API は、同じアーキテクチャで同じ数の SM を持つ GPU 上で実行される場合、毎回ビット単位で同じ結果を生成する。」

— cuBLAS Documentation (https://docs.nvidia.com/cuda/cublas/)

条件	ビット単位の再現性
同一Toolkit + 同一GPUアーキテクチャ + 同一SM数	保証される
Toolkit バージョンをまたぐ	保証されない
複数ストリームを使用	非決定的になる場合あり
GPU 世代が異なる	保証されない

科学論文の再実験や厳密な検証用途では、Toolkit バージョン・GPU 型番・ドライババージョンまで記録しておくこと。後から「数値が合わない」となっても再現できない。

10. トラブルシューティング — ライン停止時の即時対応集

エラーコード	主因	対処の第一手
`CUBLAS_STATUS_NOT_INITIALIZED`	`cublasCreate()` を呼んでいない / CUDA コンテキスト未確立	ハンドル生成位置を確認、CUDA Runtime の初期化順を見直す
`CUBLAS_STATUS_ALLOC_FAILED`	GPU メモリ不足	`nvidia-smi` で VRAM 使用量を確認、バッチサイズや行列サイズを縮小
`CUBLAS_STATUS_INVALID_VALUE`	引数が不正(サイズ・ストライド・ポインタ)	列優先である点を再確認、`lda/ldb/ldc` がそれぞれの行数以上か検査
`CUBLAS_STATUS_ARCH_MISMATCH`	使用機能が現GPUに非対応	Tensor Core 系API使用時に発生しやすい。Compute Capability を確認
`CUBLAS_STATUS_EXECUTION_FAILED`	カーネル実行中失敗	不正ポインタ・OOB アクセス・CUDA/cuBLAS バージョン不整合を疑う
`CUBLAS_STATUS_INTERNAL_ERROR`	内部エラー(ドライバ起因が多い)	NVIDIA ドライバ・CUDA Toolkit・cuBLAS のバージョン整合を見直す
`CUBLAS_STATUS_NOT_SUPPORTED`	機能未サポート	データ型/アルゴリズム/アーキ要件を確認、cuBLASLt なら別 algo を試す
「想定外に遅い」	行列サイズが8の倍数でない/FP32 のまま	サイズ調整 + 混合精度導入 + cuBLAS Logger で実行 algo を確認

PyTorch ユーザーが RuntimeError: CUDA error: CUBLAS_STATUS_NOT_INITIALIZED を見た場合、根本原因は CUDA メモリ破壊 であることが多い。スタックトレースだけ見て cuBLAS のせいにすると沼にハマる。CUDA_LAUNCH_BLOCKING=1 を付けて再実行すると本当のエラー位置が分かる。

cuBLAS の実行ログを見たければ、環境変数で出せる:

export CUBLASLT_LOG_LEVEL=5
export CUBLAS_LOGINFO_DBG=1
export CUBLAS_LOGDEST_DBG=stdout

11. 用語集 — 部品名鑑

記事中に出てきた用語を、迷ったときの辞書代わりに。

用語	一言定義
BLAS	線形代数演算の標準仕様。1979年〜
cuBLAS	NVIDIA による GPU 向け BLAS 実装
GEMM	General Matrix Multiply。`C = αAB + βC` の汎用行列積
SGEMM / DGEMM	単精度(FP32) / 倍精度(FP64) の GEMM
HGEMM	半精度(FP16) の GEMM。Tensor Core 活用の主戦場
Level 1 / 2 / 3	BLAS の階層。ベクトル × ベクトル / 行列 × ベクトル / 行列 × 行列
Tensor Core	行列乗算専用ハードウェア。Volta 世代以降
混合精度	FP16 で高速計算しつつ重要部分のみ FP32 で精度確保
エピローグ融合	GEMM 後のバイアス加算・活性化関数を1回のカーネルにまとめる
列優先	行列を列単位で連続配置する流儀(Column-major)。Fortran と cuBLAS のデフォルト
SM	Streaming Multiprocessor。NVIDIA GPU の演算ユニット単位。同じSM数のGPUでは結果が再現する

12. 学習ロードマップ

cuBLAS を起点に「どこへ進むか」を3段階で提示する。

Step 1: 入門(まずここ)

本記事を読み終えた状態
PyTorch で torch.mm を呼んで、nvidia-smi で GPU が稼働するのを確認
関連: CUDA / GPU / VRAM の基礎

Step 2: 中級(深掘り)

PyTorch の torch.cuda.amp で混合精度を試す
バッチサイズと行列サイズを変えながら、Tensor Core が効くポイントを実測
cuDNN を学び、cuBLAS との役割分担を理解
Triton で自作 GEMM カーネルを書いてみる(ベンチマーク用)

Step 3: 上級(自作領域へ)

cuBLASLt のアルゴリズム選択を手動で行う
cuBLASDx を使ったカーネル内 GEMM 融合
大規模分散学習で cuBLASXt / cuBLASMp の検討
LLM 推論最適化への応用(vLLM / llama.cpp の内部理解)

13. まとめ

工場の比喩を最後にもう一度回収しておく。

BLAS = ISO規格のような世界標準仕様(1979年〜)
cuBLAS = NVIDIA 印の最新自動化工場
GEMM = 工場で最も忙しい主要生産ライン
Tensor Core = 特殊作業を爆速でこなすロボットアーム
使うべき理由 = 最適化の深さ・世代対応・信頼性・開発効率

PyTorch / TensorFlow / Transformers を使っていれば、cuBLAS は知らずに走っている。その存在を意識するだけで、性能チューニングの解像度が一段上がる。

「行列の掛け算なんて自分で書けるよ」と思っていた頃の自分に教えてあげたいことは一つ。ガレージで車を作ろうとするな。トヨタを使え ((;ﾟдﾟ)それな)。草。

次回は、ディープラーニング専用に最適化されたライブラリ cuDNN を解説する。cuBLAS が汎用行列演算の工場なら、cuDNN は「ニューラルネット専用部品工場」だ。

シリーズ記事

第1回: CUDAってなんだ？
第2回: cuBLASってなんだ？(この記事)
第3回: cuDNNってなんだ？
ハブ記事: GPU計算の三銃士シリーズ

参考資料

公式ドキュメント

cuBLAS Documentation — cuBLAS 公式リファレンス
NVIDIA cuBLAS Product Page — 製品概要とファミリー全体像

BLAS の歴史と仕様

Netlib BLAS — BLAS 公式リファレンス実装
Wikipedia: Basic Linear Algebra Subprograms — 歴史と全体像

最適化テクニック

筆者の他の関連記事は X(旧Twitter)で随時告知している。