GPU：为什么适合 AI

GPU：为什么适合 AI

你有没有想过，一个"图形处理器"是怎么变成 AI 时代最重要的计算引擎的？这件事不是偶然，背后有非常深刻的原因。

GPU 最初的工作是每秒渲染几千万个三角形：每个三角形的颜色、光照、阴影要分别计算，这些计算之间完全独立，可以并行。这个特点和深度学习的计算需求几乎完全吻合——矩阵乘法中每个元素的计算也是独立的，可以同时进行。

Level 1：建立直觉

CPU vs GPU：两种完全不同的设计哲学

CPU（Intel i9-14900K）：
  24个核心（8 P核 + 16 E核）
  每个核心：复杂、智能
    - 乱序执行
    - 分支预测
    - 大型L1/L2缓存（每核 32KB L1 + 2MB L2）
    - 超线程
  时钟频率：5.8 GHz（最大）
  峰值算力：~1 TFLOPS（FP32）
  
GPU（NVIDIA H100 SXM）：
  16,896 个 CUDA 核心
  每个核心：简单、高效
    - 无乱序执行
    - 无分支预测
    - 极小的 L1 缓存
    - 专注做乘加
  时钟频率：1.83 GHz
  峰值算力：60 TFLOPS（FP32），~2000 TFLOPS（FP8 张量核）

CPU 是一个超级学霸：一个人能同时在脑子里处理非常复杂的逻辑。GPU 是大量普通学生的合唱团：每个人只会做简单的事情，但几万人同时工作，总产出惊人。

GPU 为什么这么快：矩阵乘法

深度学习的核心计算是矩阵乘法（GEMM）：

C[i][j] = Σ_k A[i][k] * B[k][j]

对于 1000×1000 的矩阵：
  需要计算 1000×1000 = 100万个输出元素
  每个元素需要 1000 次乘加
  总计：10亿次浮点运算

每个输出元素的计算完全独立于其他元素
→ 可以 100万 路并行！

GPU 的 16896 个核心可以同时处理 16896 个这样的乘加操作，CPU 只能处理几十个。

CUDA 编程模型的直觉

Kernel（GPU 函数）：在 GPU 上并行执行的函数：

// GPU Kernel：向量加法
__global__ void vector_add(float* A, float* B, float* C, int N) {
    int idx = blockIdx.x * blockDim.x + threadIdx.x;
    if (idx < N) {
        C[idx] = A[idx] + B[idx];  // 每个线程处理一个元素
    }
}

// 主机（CPU）代码
int N = 1000000;
// 配置：1000个块，每块1024个线程 = 1024000个线程同时运行
vector_add<<<1000, 1024>>>(d_A, d_B, d_C, N);

一次函数调用，100万个线程同时运行，每个只做一次加法——这就是 GPU 的并行哲学。

为什么神经网络天然适合 GPU

神经网络的前向传播是一系列矩阵乘法：

输入 X (batch × in_features)
× 权重 W (in_features × out_features)
+ 偏置 b
→ 输出 Y (batch × out_features)

每层就是一个大矩阵乘法
每个输出元素独立计算
= GPU 的完美应用场景

GPT-3（175B 参数）的前向传播：
  约 96 层 Transformer
  每层包含多个矩阵乘法
  单次推理（512 token）：约 350 万亿次浮点运算
  → GPU 在几秒内完成，CPU 需要几小时

Level 2：原理剖析

GPU 架构深探

NVIDIA GPU 的层级结构：

GPU
├── GPC（Graphics Processing Cluster）× 7
│   └── SM（Streaming Multiprocessor）× 132 (H100 SXM)
│       ├── CUDA 核心 × 128（整数+浮点）
│       ├── 张量核（Tensor Core）× 4（矩阵加速）
│       ├── 寄存器文件：65,536 个 32-bit 寄存器
│       ├── L1/共享内存：228 KB
│       └── Warp 调度器 × 4
│
├── L2 缓存：50 MB（H100）
└── HBM3 显存：80 GB，3.35 TB/s

Warp（线程束）：GPU 调度的基本单位：

Warp = 32 个线程，总是一起执行同一条指令（SIMT）

SM 内同时运行：最多 64 个 warp = 2048 个线程
H100 共 132 个 SM × 2048 = 270,336 个并发线程

但！如果 32 个线程走了不同的分支（Branch Divergence）：
    线程0-15：if 分支
    线程16-31：else 分支
→ 两个分支顺序执行，性能减半！
→ GPU 代码要尽量避免分支

张量核（Tensor Core）：AI 加速的秘密

普通 CUDA 核心：每周期处理 1 次 FP32 乘加。

张量核（Tensor Core）：每周期处理一个 4×4 矩阵乘法：

张量核操作（每时钟周期）：
D[4×4] = A[4×4] × B[4×4] + C[4×4]
= 16×16 = 256 次乘加（FP16）

H100 有 528 个 Tensor Core（132 SM × 4）
528 × 256 = 135,168 次/周期（FP16）
× 1.83 GHz = 约 250 TFLOPS（FP16）

NVIDIA 每代都在升级张量核：Volta（2017）→ Turing → Ampere → Hopper（H100）

张量核精度支持：
FP64: ~60 TFLOPS（H100）
FP32: ~60 TFLOPS
BF16: ~2000 TFLOPS
FP8:  ~4000 TFLOPS（H100 SXM）

越低精度 → 越多算力（AI 训练用 BF16/FP8，推理可以更低）

显存带宽：另一个瓶颈

GPU 的计算能力惊人，但显存带宽常常是瓶颈：

H100 SXM：
  算力：60 TFLOPS（FP32）
  显存带宽：3.35 TB/s
  
  算术强度（Arithmetic Intensity）= 操作数 / 内存字节数
  
  向量加法：1次操作，读2字节+写1字节 → 强度 = 0.33 FLOP/byte
    → Memory-Bound：受 3.35 TB/s 限制
    → 实际 TFLOPS 远低于峰值
    
  矩阵乘法（N=10000）：N³/N² = N FLOP/byte → Memory-Bound 边界
    → 大矩阵：Compute-Bound，接近峰值算力
    → 小矩阵/推理（batch=1）：Memory-Bound，利用率低

这就是为什么 AI 推理加速如此困难——当 batch size 小时，GPU 算力大部分在等内存。

CUDA/显存/CPU 内存的通信

通信路径：
CPU → [PCIe 4.0 x16, 64 GB/s] → GPU 显存（HBM）
GPU 显存 → [NVLink, 900 GB/s] → 另一张 GPU 显存（H100 NVLink 带宽）

多 GPU 训练的数据流：
  梯度同步：各 GPU 计算完梯度 → 通过 NVLink AllReduce → 更新所有 GPU 的参数

PCIe 带宽（64 GB/s）远低于 HBM 带宽（3.35 TB/s），这意味着：

• CPU-GPU 数据传输是瓶颈
• 尽量在 GPU 显存里做所有操作（减少 cudaMemcpy）
• PyTorch、TensorFlow 的 DataLoader 要用 pinned memory 加速传输

Level 3 · 规范怎么定义的（资深）

GPU 计算的编程模型与标准

GPU 计算的编程模型由多个标准和框架定义。CUDA（Compute Unified Device Architecture）是 NVIDIA 的专有框架，其编程模型在 CUDA C++ Programming Guide 中定义：程序由在 GPU 上并行执行的 kernel 函数组成，线程被组织为 Grid → Block → Thread 的三级层次结构。每个 Block 内的线程共享 Shared Memory（用户管理的高速片上 SRAM，延迟约 20-30 个周期），通过 __syncthreads() 同步。CUDA 的内存模型定义了五种内存空间：寄存器（每线程）、Shared Memory（每 Block）、L1/L2 Cache、Global Memory（全局 DRAM，延迟约 400-800 个周期）和 Constant Memory。

OpenCL（Open Computing Language，由 Khronos Group 维护）是跨平台的 GPU 计算标准（OpenCL 3.0，2020 年），支持 NVIDIA、AMD、Intel 和 ARM 的 GPU。OpenCL 的执行模型与 CUDA 类似（Work-group ≈ Block，Work-item ≈ Thread），但 OpenCL 是规范定义的标准，而 CUDA 是实现定义的框架。

张量核（Tensor Core）的行为在 NVIDIA 的 PTX ISA 文档中定义。Tensor Core 执行的基本操作是 D = A × B + C，其中 A、B、C、D 是小矩阵（如 16×16×16 的 FP16 矩阵）。Hopper 架构（H100）的 Tensor Core 支持 FP64、TF32、FP16、BF16、INT8 和 FP8 精度。FP8（E4M3 和 E5M2 两种格式）的规范由 NVIDIA、ARM 和 Intel 在 2022 年联合发布的白皮书中定义，是 AI 训练和推理中精度与性能的最新权衡点。

Level 4 · 边界与陷阱（所有人）

陷阱 1：GPU 利用率低的"内存带宽瓶颈"

GPU 有数千个计算核心，但如果数据从显存（VRAM）到计算核心的带宽不足，大部分核心会处于空闲等待状态。H100 的 HBM3 带宽为 3.35 TB/s，但一个简单的向量加法 C[i] = A[i] + B[i] 只做了 1 次加法却需要读 2 个浮点数、写 1 个浮点数（12 字节/元素），计算/访存比（Operational Intensity）只有 1/12 FLOP/Byte。H100 的峰值算力约 1000 TFLOPS（FP16），但受限于 3.35 TB/s 的带宽，实际只能支持约 3.35 TFLOPS 的向量加法——利用率仅 0.3%。这就是为什么 AI 框架要用算子融合（Operator Fusion）将多个操作合并，增加每次显存访问的计算量。

陷阱 2：GPU 的 Warp Divergence 导致实际并行度减半

NVIDIA GPU 以 32 个线程为一组（称为 Warp）同时执行同一条指令。如果 Warp 内的线程在 if-else 分支处走了不同路径，GPU 必须先执行 if 分支（不走这条路径的线程闲置），再执行 else 分支（反过来闲置）——这就是 Warp Divergence。最坏情况下，一个 if-else 分支会让执行时间翻倍。深度学习中的 Dropout 层、条件计算（Mixture-of-Experts）和稀疏注意力机制都容易触发 Warp Divergence。NVIDIA 的 Nsight Compute 工具可以量化 Divergence 的影响。

陷阱 3：CPU-GPU 数据传输是隐藏的性能杀手

PCIe 4.0 x16 的带宽约 32 GB/s（双向），而 GPU 内部的 HBM 带宽可达 3000+ GB/s——差距约 100 倍。如果你的代码频繁在 CPU 和 GPU 之间传输数据（如每个 batch 都把数据从 CPU 内存拷贝到 GPU 显存），传输时间可能超过计算时间。PyTorch 的 DataLoader 使用 pin_memory=True 来将数据放在"锁页内存"（pinned memory）中，避免 DMA 传输前的额外拷贝；non_blocking=True 参数让数据传输与 GPU 计算重叠。更进一步，NVIDIA 的 GPUDirect RDMA 允许网卡直接向 GPU 显存传输数据，完全绕过 CPU——这是分布式训练的关键优化。