TMA (Tensor Memory Accelerator・Hopper)とは？（ティーエムエー）意味・特徴をわかりやすく解説

Q: TMA (Tensor Memory Accelerator・Hopper)とは？

NVIDIA Hopper (H100・2022) で導入のメモリ転送専用アクセラレータ。GPU SM ↔ Shared Memory 間の大規模Tensor 転送を非同期・低消費電力で実行。

概要

TMA (Tensor Memory Accelerator) はNVIDIA Hopper アーキテクチャ (H100・GTC2022 March・2022年3月発表) で導入されたメモリ転送専用ハードウェアアクセラレータ。従来Warp (32 threads) 単位で実行していたShared Memory ↔ Global Memory 転送を、TMA が SM (Streaming Multiprocessor) 全体で1命令発火・SM 計算リソース解放しながら非同期実行する設計。AI 学習 / 推論のメモリ帯域ボトルネックを大幅軽減する。

TMA の役割

従来 (Ampere まで)	TMA (Hopper以降)
Warp 32 threads 同期コピー	TMA 1命令発火
SM ALU 占有	SM 解放可 (非同期)
1サイクル 128 bytes	数MB 1命令
プログラム複雑	API 単純化

TMA 機能

Global Memory → Shared Memory 非同期コピー
Shared Memory → Global Memory 非同期コピー
Tensor Stride (多次元) 対応
最大次元: 5D テンソル
データタイプ: 1-128 bytes/element
容量: 数MB / 命令まで
メモリポリシ: Cache hint 指定可

CUDA API (Hopper)

// cuda::barrier 同期
__shared__ alignas(16) char buffer[128 * 1024];
__shared__ cuda::barrier&lt;cuda::thread_scope_block> bar;

if (threadIdx.x == 0) {
    init(&amp;bar, blockDim.x);
}
__syncthreads();

cuda::memcpy_async(global_ptr, shared_buffer, bar);
bar.arrive_and_wait();

cuda::memcpy_async API がTMA を内部使用、開発者は同期barrier のみ気にすれば良い。

AI 学習 / 推論への効果

ワークロード	Ampere	Hopper (TMA有)	改善
GPT-3 175B 学習	1.0×	2.0×	2倍
Llama 70B 推論	1.0×	1.8×	1.8倍
BERT-Large	1.0×	1.5×	1.5倍

TMA + Tensor Core Gen4 + Transformer Engine の組み合わせでAmpere比2倍超の学習速度。

TMA 2.0 (Blackwell・2024)

NVIDIA Blackwell (RTX 50 / B100 / B200・2024年) ではTMA 2.0 として強化:

Multi-Cast 機能: 1 メモリ転送を複数SM へ並列送信
Cluster-wide TMA: GPC (Graphics Processing Cluster) 単位の協調
Tensor Core Gen5 と統合・FP4 / FP8 直接フィード
帯域効率 30-50% 改善

主要対応GPU

GPU	TMA
H100 (SXM5 / PCIe)	TMA 1.0
H200	TMA 1.0 (HBM3E)
B100 / B200	TMA 2.0
RTX 50 (Blackwell コンシューマ)	TMA 2.0 (限定機能)
GH200 (Grace Hopper)	TMA 1.0
GB200 NVL72	TMA 2.0

FAQ

Q1: 開発者は意識する? A: 直接TMA 命令使用は通常不要。CUDA 12+ の cuda::memcpy_async / cuTensorMapEncodeTiled 等が内部利用、cuBLAS / cuDNN も透過的TMA 活用。

Q2: TMA がない世代でも動く? A: Volta / Ampere は対応API でフォールバック (Warp 単位コピー) 実行。性能はTMA なしの方が遅いがコード互換性維持。

Q3: AI 以外の用途? A: 科学技術計算 (CFD / FEM / FFT) でも大規模Tensor の頻繁な転送が発生・恩恵あり。一般CUDA カーネル開発でも次第に重要視。

TMA (Tensor Memory Accelerator・Hopper)（ティーエムエー）

概要