CUDA 编程：cuBLAS SGEMM 与手写矩阵乘法性能对比

文章目录

• 1. 背景
• 2. 四种矩阵乘法实现
  • 2.1 CPU MatMul
  • 2.2 CUDA Naive MatMul
  • 2.3 Shared Memory Tiled MatMul
  • 2.4 cuBLAS SGEMM
• 3. 实验环境
• 4. 实验结果
  • 4.1 执行时间
  • 4.2 GFLOPS 对比
  • 4.3 加速效果
• 5. 结果分析
  • 5.1 Shared Memory 为什么比 Naive 快？
  • 5.2 为什么 cuBLAS 快这么多？
    • Register Blocking
    • 多级 Tile
    • 更优秀的访存策略
• 6. 实验结论

从手写 CUDA Kernel 到工业级高性能库，理解 GPU 矩阵计算优化的差距。

1. 背景

矩阵乘法（GEMM, General Matrix Multiplication）是 GPU 计算中最重要的算子之一。

在深度学习中：

• Transformer 的 QKV 投影
• Linear 层
• Attention 计算
• MLP 网络

都大量依赖矩阵乘法。

本节通过实现并比较：

1. CPU 矩阵乘法
2. CUDA Naive Kernel
3. CUDA Shared Memory Tiled Kernel
4. NVIDIA cuBLAS SGEMM

理解 GPU 矩阵计算的优化过程。

在这里插入图片描述

在这里插入图片描述

2. 四种矩阵乘法实现

2.1 CPU MatMul

最基础实现：

for row
    for col
        sum = 0
        for k
            C[row][col]+=A[row][k]*B[k][col]

特点：

• 逻辑简单
• 单线程执行
• 计算能力有限

2.2 CUDA Naive MatMul

GPU 中：

一个线程计算 C 中一个元素

结构：

Thread(0,0) → C[0][0]

Thread(0,1) → C[0][1]

Thread(1,0) → C[1][0]

优点：

• 并行度高
• 容易实现

问题：

每个线程都会重复读取：

Global Memory
      ↓
      A
      B

大量访存浪费。

2.3 Shared Memory Tiled MatMul

优化思路：

利用 Shared Memory 保存矩阵块。

计算过程：

Global Memory

      ↓

Tile加载

      ↓

Shared Memory

      ↓

多个线程重复使用

      ↓

计算结果

优势：

• 减少 Global Memory 访问
• 提高数据复用率
• 降低访存瓶颈

2.4 cuBLAS SGEMM

cuBLAS 是 NVIDIA 官方高性能线性代数库。

调用：

cublasSgemm()

内部包含大量优化：

• 多级 Tile
• Register Blocking
• Warp级优化
• Memory访问优化
• Tensor Core支持（部分场景）

因此性能远高于普通 CUDA Kernel。

3. 实验环境

GPU：

Tesla T4

测试：

矩阵：
512×512
1024×1024
2048×2048

数据类型：
FP32

重复次数：
512/1024: 10次
2048: 3次

计时范围：

仅计算时间

不包含：
cudaMalloc
cudaMemcpy
初始化

4. 实验结果

4.1 执行时间

可以看到：

随着矩阵规模增加：

CPU差距快速扩大

cuBLAS优势越来越明显

4.2 GFLOPS 对比

性能趋势：

CPU < CUDA Naive < Shared Memory < cuBLAS

4.3 加速效果

在这里插入图片描述

在这里插入图片描述

5. 结果分析

5.1 Shared Memory 为什么比 Naive 快？

Naive：

Thread

 ↓

Global Memory

 ↓

读取A/B

 ↓

计算

Shared：

Global Memory

 ↓

一次加载Tile

 ↓

Shared Memory缓存

 ↓

多个线程复用

减少了大量：

Global Memory访问次数

因此：

Shared > Naive

实验：

2048×2048：

Naive:
61.3 ms

Shared:
25.5 ms

提升：

2.4倍

5.2 为什么 cuBLAS 快这么多？

Shared Memory Kernel：

主要优化：

Global Memory
        ↓
Shared Memory
        ↓
计算

而 cuBLAS：

Global Memory

 ↓

多级Tile

 ↓

Shared Memory

 ↓

Register

 ↓

Warp级计算

 ↓

Tensor Core/专用硬件

它进一步优化：

Register Blocking

普通 Kernel：

一个线程计算一个C元素

高性能 GEMM：

一个线程计算多个C元素

更多计算放在 Register 中。

多级 Tile

普通：

Block Tile

cuBLAS：

Block Tile

    ↓

Warp Tile

    ↓

Thread Tile

    ↓

Register Tile

充分利用 GPU 层次结构。

更优秀的访存策略

包括：

• Coalesced Memory Access
• Shared Memory Bank Conflict优化
• Cache利用
• Prefetch

6. 实验结论

本实验得到：

1. GPU远快于CPU
2. Shared Memory 是有效优化
3. cuBLAS 是工程首选，实际项目中不要重复造轮子，矩阵计算优先使用 cuBLAS / CUTLASS。

在这里插入图片描述

在这里插入图片描述

一句话总结：

手写 CUDA Kernel 帮助理解 GPU 原理，而 cuBLAS 代表 NVIDIA 工程团队多年优化后的极限性能。实际 AI 推理和训练系统中，矩阵计算通常直接依赖 cuBLAS、cuDNN 或 CUTLASS。