01-GPU 架构概述¶
重要性：⭐⭐⭐⭐⭐ 实用度：⭐⭐⭐⭐⭐ 学习时间： 2 天 必须掌握：是
为什么学这一章¶
GPU 已经从单纯的图形处理器发展成为通用并行计算设备。理解 GPU 架构能帮助你： - 理解为什么 GPU 适合 AI 和深度学习 - 编写高效的并行程序 - 优化 CUDA 程序性能 - 理解现代 AI 框架的底层实现
学完这一章，你将能够： - ✅ 解释 GPU 与 CPU 的设计差异 - ✅ 理解 GPU 的核心架构组件 - ✅ 掌握 GPU 编程的基本概念 - ✅ 了解 GPU 的应用场景
📖 核心概念¶
1. GPU 与 CPU 的设计哲学¶
Text Only
┌─────────────────────────────────────────────────────────────────────┐
│ CPU vs GPU 设计哲学对比 │
├─────────────────────────────────────────────────────────────────────┤
│ │
│ CPU（中央处理器） │
│ ┌───────────────────────────────────────────────────────────────┐ │
│ │ 设计目标：最小化单个任务的延迟（Latency） │ │
│ │ │ │
│ │ 特点： │ │
│ │ • 强大的单核性能 │ │
│ │ • 复杂的控制逻辑（分支预测、乱序执行） │ │
│ │ • 大容量缓存（减少内存访问延迟） │ │
│ │ • 适合串行任务 │ │
│ │ │ │
│ │ 比喻：F1赛车 - 速度快但昂贵，适合复杂赛道 │ │
│ │ │ │
│ │ ┌─────┐ ┌─────┐ ┌─────┐ ┌─────┐ │ │
│ │ │核心1│ │核心2│ │核心3│ │核心4│ ← 少量强大核心 │ │
│ │ └──┬──┘ └──┬──┘ └──┬──┘ └──┬──┘ │ │
│ │ └───────┴───────┴───────┘ │ │
│ │ 共享L3缓存 │ │
│ └───────────────────────────────────────────────────────────────┘ │
│ │
│ GPU（图形处理器） │
│ ┌───────────────────────────────────────────────────────────────┐ │
│ │ 设计目标：最大化吞吐量（Throughput） │ │
│ │ │ │
│ │ 特点： │ │
│ │ • 大量简单核心（数千个） │ │
│ │ • 简单的控制逻辑（无分支预测、SIMT顺序流水线） │ │
│ │ • 小容量缓存（依赖高带宽内存） │ │
│ │ • 适合并行任务 │ │
│ │ │ │
│ │ 比喻：货运船队 - 单船慢但数量多，适合运输大量货物 │ │
│ │ │ │
│ │ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ │ │
│ │ │核│ │核│ │核│ │核│ │核│ │核│ │核│ │核│ ← 大量简单核心 │ │
│ │ └──┘ └──┘ └──┘ └──┘ └──┘ └──┘ └──┘ └──┘ │ │
│ │ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ ┌──┐ │ │
│ │ │核│ │核│ │核│ │核│ │核│ │核│ │核│ │核│ │ │
│ │ └──┘ └──┘ └──┘ └──┘ └──┘ └──┘ └──┘ └──┘ │ │
│ │ ... 数千个核心 ... │ │
│ └───────────────────────────────────────────────────────────────┘ │
│ │
└─────────────────────────────────────────────────────────────────────┘
核心差异对比¶
| 特性 | CPU | GPU |
|---|---|---|
| 核心数量 | 4-64 个 | 数千个 |
| 核心复杂度 | 复杂（乱序执行、分支预测） | 简单（ SIMT 模型，同一 Warp 内线程同步执行相同指令） |
| 时钟频率 | 3-5 GHz | 1-2 GHz |
| 缓存大小 | 大（ MB 级） | 小（ KB 级） |
| 内存带宽 | 50-100 GB/s | 500-1000+ GB/s |
| 设计目标 | 低延迟 | 高吞吐量 |
| 适用任务 | 串行、复杂逻辑 | 并行、计算密集 |
2. GPU 架构详解¶
NVIDIA GPU 架构（以 Ampere 为例）¶
Text Only
┌─────────────────────────────────────────────────────────────────────┐
│ NVIDIA GPU架构（Ampere） │
├─────────────────────────────────────────────────────────────────────┤
│ │
│ GPU（图形处理器） │
│ ┌───────────────────────────────────────────────────────────────┐ │
│ │ │ │
│ │ ┌─────────────┐ ┌─────────────┐ ┌─────────────┐ │ │
│ │ │ GPC 0 │ │ GPC 1 │ │ GPC 2... │ ← 图形处理集群│ │
│ │ │ (Graphics │ │ │ │ │ │ │
│ │ │ Processing │ │ │ │ │ │ │
│ │ │ Cluster) │ │ │ │ │ │ │
│ │ └──────┬──────┘ └──────┬──────┘ └──────┬──────┘ │ │
│ │ │ │ │ │ │
│ │ ┌──────┴──────┐ ┌──────┴──────┐ ┌──────┴──────┐ │ │
│ │ │ TPC 0 │ │ TPC 0 │ │ TPC 0 │ ← 纹理处理集群│ │
│ │ │ TPC 1 │ │ TPC 1 │ │ TPC 1 │ │ │
│ │ └──────┬──────┘ └──────┬──────┘ └──────┬──────┘ │ │
│ │ │ │ │ │ │
│ │ ┌──────┴──────┐ ┌──────┴──────┐ ┌──────┴──────┐ │ │
│ │ │ SM 0 │ │ SM 0 │ │ SM 0 │ ← 流式多处理器│ │
│ │ │ SM 1 │ │ SM 1 │ │ SM 1 │ (Streaming │ │
│ │ │ ... │ │ ... │ │ ... │ Multiprocessor)│
│ │ └──────┬──────┘ └──────┬──────┘ └──────┬──────┘ │ │
│ │ │ │ │ │ │
│ │ ┌──────┴──────┐ ┌──────┴──────┐ ┌──────┴──────┐ │ │
│ │ │ CUDA Core │ │ CUDA Core │ │ CUDA Core │ ← CUDA核心 │ │
│ │ │ Tensor Core │ │ Tensor Core │ │ Tensor Core │ ← Tensor核心 │ │
│ │ │ RT Core │ │ RT Core │ │ RT Core │ ← 光追核心 │ │
│ │ │ ... │ │ ... │ │ ... │ │ │
│ │ └─────────────┘ └─────────────┘ └─────────────┘ │ │
│ │ │ │
│ │ ┌─────────────────────────────────────────────────────────┐ │ │
│ │ │ 全局内存（Global Memory） │ │ │
│ │ │ 容量：数GB到数十GB │ │ │
│ │ │ 带宽：数百GB/s到TB/s │ │ │
│ │ └─────────────────────────────────────────────────────────┘ │ │
│ │ │ │
│ └───────────────────────────────────────────────────────────────┘ │
│ │
└─────────────────────────────────────────────────────────────────────┘
流式多处理器（ SM ）详解¶
Text Only
┌─────────────────────────────────────────────────────────────────────┐
│ 流式多处理器（SM）内部结构 │
├─────────────────────────────────────────────────────────────────────┤
│ │
│ SM（Streaming Multiprocessor） │
│ ┌───────────────────────────────────────────────────────────────┐ │
│ │ │ │
│ │ ┌─────────────┐ ┌─────────────┐ ┌─────────────┐ │ │
│ │ │ Warp调度器1 │ │ Warp调度器2 │ │ Warp调度器3 │ ← 调度器 │ │
│ │ │ Warp调度器4 │ │ │ │ │ │ │
│ │ └──────┬──────┘ └──────┬──────┘ └──────┬──────┘ │ │
│ │ │ │ │ │ │
│ │ ┌──────┴──────┐ ┌──────┴──────┐ ┌──────┴──────┐ │ │
│ │ │ CUDA Core │ │ CUDA Core │ │ CUDA Core │ ← FP32单元│ │
│ │ │ (FP32) │ │ (FP32) │ │ (FP32) │ │ │
│ │ │ CUDA Core │ │ CUDA Core │ │ CUDA Core │ │ │
│ │ │ (FP32) │ │ (FP32) │ │ (FP32) │ │ │
│ │ │ ... │ │ ... │ │ ... │ │ │
│ │ │ (64-128个) │ │ │ │ │ │ │
│ │ └─────────────┘ └─────────────┘ └─────────────┘ │ │
│ │ │ │
│ │ ┌─────────────┐ ┌─────────────┐ ┌─────────────┐ │ │
│ │ │ Tensor Core │ │ Tensor Core │ │ Tensor Core │ ← Tensor核心│ │
│ │ │ (4-8个) │ │ │ │ │ (AI加速) │ │
│ │ └─────────────┘ └─────────────┘ └─────────────┘ │ │
│ │ │ │
│ │ ┌─────────────────────────────────────────────────────────┐ │ │
│ │ │ 共享内存（Shared Memory） │ │ │
│ │ │ 容量：几十KB到几百KB │ │ │
│ │ │ 速度：比全局内存快100倍 │ │ │
│ │ └─────────────────────────────────────────────────────────┘ │ │
│ │ │ │
│ │ ┌─────────────────────────────────────────────────────────┐ │ │
│ │ │ 寄存器文件（Register File） │ │ │
│ │ │ 容量：数百KB │ │ │
│ │ │ 每个线程有私有寄存器 │ │ │
│ │ └─────────────────────────────────────────────────────────┘ │ │
│ │ │ │
│ └───────────────────────────────────────────────────────────────┘ │
│ │
└─────────────────────────────────────────────────────────────────────┘
3. GPU 编程模型¶
CUDA 线程层次结构¶
Text Only
┌─────────────────────────────────────────────────────────────────────┐
│ CUDA线程层次结构 │
├─────────────────────────────────────────────────────────────────────┤
│ │
│ Grid（网格）- 整个GPU程序 │
│ ┌───────────────────────────────────────────────────────────────┐ │
│ │ │ │
│ │ Block(0,0) Block(1,0) Block(2,0) │ │
│ │ ┌─────────┐ ┌─────────┐ ┌─────────┐ │ │
│ │ │Thread │ │Thread │ │Thread │ │ │
│ │ │(0,0) │ │(0,0) │ │(0,0) │ │ │
│ │ │(1,0) │ │(1,0) │ │(1,0) │ │ │
│ │ │(2,0) │ │(2,0) │ │(2,0) │ │ │
│ │ │... │ │... │ │... │ │ │
│ │ │(0,1) │ │(0,1) │ │(0,1) │ │ │
│ │ │... │ │... │ │... │ │ │
│ │ └─────────┘ └─────────┘ └─────────┘ │ │
│ │ │ │
│ │ Block(0,1) Block(1,1) Block(2,1) │ │
│ │ ┌─────────┐ ┌─────────┐ ┌─────────┐ │ │
│ │ │Thread │ │Thread │ │Thread │ │ │
│ │ │... │ │... │ │... │ │ │
│ │ └─────────┘ └─────────┘ └─────────┘ │ │
│ │ │ │
│ │ ... │ │
│ │ │ │
│ └───────────────────────────────────────────────────────────────┘ │
│ │
│ 层次关系： │
│ ├── Grid：包含多个Block，可以是一维、二维或三维 │
│ ├── Block：包含多个Thread，可以是一维、二维或三维 │
│ ├── Thread：最基本的执行单元 │
│ └── Warp：32个Thread组成一个Warp，是调度的基本单位 │
│ │
│ 示例： │
│ dim3 gridDim(3, 2); // 3x2 = 6个Block │
│ dim3 blockDim(4, 4); // 4x4 = 16个Thread/Block │
│ 总线程数 = 6 * 16 = 96个线程 │
│ │
└─────────────────────────────────────────────────────────────────────┘
线程索引计算¶
C++
// CUDA内核函数示例
__global__ void vectorAdd(float* A, float* B, float* C, int n) {
// 计算全局线程ID
int i = blockIdx.x * blockDim.x + threadIdx.x;
// 每个线程处理一个元素
if (i < n) {
C[i] = A[i] + B[i];
}
}
// 调用内核
int main() {
int n = 1000000;
// 配置：256线程/Block，足够多的Block覆盖所有数据
int threadsPerBlock = 256;
int blocksPerGrid = (n + threadsPerBlock - 1) / threadsPerBlock;
vectorAdd<<<blocksPerGrid, threadsPerBlock>>>(d_A, d_B, d_C, n);
return 0;
}
4. GPU 内存层次结构¶
Text Only
┌─────────────────────────────────────────────────────────────────────┐
│ GPU内存层次结构 │
├─────────────────────────────────────────────────────────────────────┤
│ │
│ 速度 ↑ │
│ 快 │ ┌───────────────────────────────────────────────────┐ │
│ │ │ 寄存器（Register） │ │
│ │ │ • 每个线程私有 │ │
│ │ │ • 容量：几十KB/SM │ │
│ │ │ • 延迟：1周期 │ │
│ │ │ • 速度最快 │ │
│ │ └───────────────────────────────────────────────────┘ │
│ │ ↓ │
│ │ ┌───────────────────────────────────────────────────┐ │
│ │ │ 共享内存（Shared Memory / L1 Cache） │ │
│ │ │ • Block内所有线程共享 │ │
│ │ │ • 容量：几十KB到几百KB │ │
│ │ │ • 延迟：20-30周期 │ │
│ │ │ • 可编程管理 │ │
│ │ └───────────────────────────────────────────────────┘ │
│ │ ↓ │
│ │ ┌───────────────────────────────────────────────────┐ │
│ │ │ L2缓存 │ │
│ │ │ • 所有SM共享 │ │
│ │ │ • 容量：几MB │ │
│ │ │ • 延迟：几百周期 │ │
│ │ └───────────────────────────────────────────────────┘ │
│ │ ↓ │
│ 慢 │ ┌───────────────────────────────────────────────────┐ │
│ │ │ 全局内存（Global Memory / 显存） │ │
│ │ │ • 所有线程可访问 │ │
│ │ │ • 容量：几GB到几十GB │ │
│ │ │ • 延迟：几百到几千周期 │ │
│ │ │ • 带宽：数百GB/s │ │
│ │ └───────────────────────────────────────────────────┘ │
│ └ │
│ 容量 → │
│ 小 大 │
│ │
└─────────────────────────────────────────────────────────────────────┘
内存类型对比¶
| 内存类型 | 作用域 | 生命周期 | 速度 | 容量 | 用途 |
|---|---|---|---|---|---|
| 寄存器 | 线程 | 线程 | 最快 | 有限 | 局部变量 |
| 共享内存 | Block | Block | 很快 | 几十 KB | 线程间通信 |
| L2 缓存 | 所有线程 | 程序 | 快 | 几 MB | 自动缓存 |
| 全局内存 | 所有线程 | 程序 | 慢 | 几 GB | 主数据存储 |
| 常量内存 | 所有线程 | 程序 | 快（缓存） | 64KB | 只读常量 |
| 纹理内存 | 所有线程 | 程序 | 快（缓存） | 大 | 图像处理 |
🧪 动手实验¶
实验 1 ：查看 GPU 信息¶
目的：了解你的 GPU 硬件信息
步骤：
- 安装 CUDA 工具包（ Linux ）
- 使用 nvidia-smi 查看 GPU 信息
- 使用 deviceQuery 查看详细信息
观察输出： - GPU 型号和架构 - CUDA 核心数量 - 显存大小和带宽 - 计算能力（ Compute Capability ）
实验 2 ：第一个 CUDA 程序¶
目的：编写并运行简单的 CUDA 程序
步骤：
- 创建 CUDA 文件
C++
// first_cuda.cu
#include <iostream>
#include <cuda_runtime.h>
// CUDA内核函数
__global__ void helloFromGPU() {
int tid = blockIdx.x * blockDim.x + threadIdx.x;
printf("Hello from thread %d (Block %d, Thread %d)\n",
tid, blockIdx.x, threadIdx.x);
}
int main() {
std::cout << "Hello from CPU!" << std::endl;
// 启动内核：2个Block，每个Block 4个线程
helloFromGPU<<<2, 4>>>();
// 等待GPU完成
cudaDeviceSynchronize();
std::cout << "Done!" << std::endl;
return 0;
}
- 编译运行
- 观察输出
- 总共 8 个线程（ 2 Block × 4 Thread ）
- 每个线程打印自己的 ID
实验 3 ：向量加法¶
目的：实现并优化向量加法
步骤：
- 创建 CUDA 文件
C++
// vector_add.cu
#include <iostream> // 引入头文件
#include <cuda_runtime.h>
#include <chrono>
// CUDA内核：向量加法
__global__ void vectorAdd(const float* A, const float* B, float* C, int n) { // 指针：存储变量的内存地址
int i = blockIdx.x * blockDim.x + threadIdx.x;
if (i < n) {
C[i] = A[i] + B[i];
}
}
// CPU版本：用于对比
void vectorAddCPU(const float* A, const float* B, float* C, int n) {
for (int i = 0; i < n; i++) {
C[i] = A[i] + B[i];
}
}
int main() {
int n = 10000000; // 1000万元素
size_t size = n * sizeof(float);
// 分配主机内存
float* h_A = new float[n];
float* h_B = new float[n];
float* h_C = new float[n];
float* h_C_CPU = new float[n];
// 初始化数据
for (int i = 0; i < n; i++) {
h_A[i] = rand() / (float)RAND_MAX;
h_B[i] = rand() / (float)RAND_MAX;
}
// CPU计算
auto start = std::chrono::high_resolution_clock::now();
vectorAddCPU(h_A, h_B, h_C_CPU, n);
auto end = std::chrono::high_resolution_clock::now();
auto cpu_time = std::chrono::duration_cast<std::chrono::microseconds>(end - start).count();
// 分配设备内存
float *d_A, *d_B, *d_C;
cudaMalloc(&d_A, size);
cudaMalloc(&d_B, size);
cudaMalloc(&d_C, size);
// 拷贝数据到设备
cudaMemcpy(d_A, h_A, size, cudaMemcpyHostToDevice);
cudaMemcpy(d_B, h_B, size, cudaMemcpyHostToDevice);
// 启动内核
int threadsPerBlock = 256;
int blocksPerGrid = (n + threadsPerBlock - 1) / threadsPerBlock;
cudaEvent_t start_gpu, stop_gpu;
cudaEventCreate(&start_gpu);
cudaEventCreate(&stop_gpu);
cudaEventRecord(start_gpu);
vectorAdd<<<blocksPerGrid, threadsPerBlock>>>(d_A, d_B, d_C, n);
cudaEventRecord(stop_gpu);
cudaEventSynchronize(stop_gpu);
float gpu_time = 0;
cudaEventElapsedTime(&gpu_time, start_gpu, stop_gpu);
// 拷贝结果回主机
cudaMemcpy(h_C, d_C, size, cudaMemcpyDeviceToHost);
// 验证结果
bool correct = true;
for (int i = 0; i < n; i++) {
if (fabs(h_C[i] - h_C_CPU[i]) > 1e-5) {
correct = false;
break;
}
}
std::cout << "CPU Time: " << cpu_time / 1000.0 << " ms" << std::endl;
std::cout << "GPU Time: " << gpu_time << " ms" << std::endl;
std::cout << "Speedup: " << cpu_time / 1000.0 / gpu_time << "x" << std::endl;
std::cout << "Result: " << (correct ? "CORRECT" : "INCORRECT") << std::endl;
// 清理
delete[] h_A; delete[] h_B; delete[] h_C; delete[] h_C_CPU;
cudaFree(d_A); cudaFree(d_B); cudaFree(d_C);
return 0;
}
- 编译运行
- 观察加速比
- 通常 GPU 比 CPU 快 10-100 倍（取决于 GPU 型号和数据大小）
💡 核心要点总结¶
GPU vs CPU¶
| 特性 | CPU | GPU |
|---|---|---|
| 核心数 | 少（ 4-64 ） | 多（数千） |
| 核心复杂度 | 高 | 低 |
| 设计目标 | 低延迟 | 高吞吐量 |
| 适用场景 | 串行任务 | 并行任务 |
GPU 架构层次¶
Text Only
GPU
├── GPC（图形处理集群）
│ └── TPC（纹理处理集群）
│ └── SM（流式多处理器）
│ ├── CUDA Core（计算核心）
│ ├── Tensor Core（AI加速）
│ ├── 共享内存
│ └── 寄存器
└── 全局内存
CUDA 线程层次¶
GPU 内存层次¶
❓ 常见问题¶
Q1 ：为什么 GPU 适合深度学习？
A ：深度学习主要是矩阵运算，具有： - 高并行性：矩阵乘法可以分解成大量独立的乘加运算 - 计算密集：需要大量浮点运算 - 数据局部性：可以充分利用缓存和共享内存
Q2 ：所有程序都适合用 GPU 加速吗？
A ：不是。适合 GPU 的程序特征： - 大规模数据并行 - 计算密集（计算/访存比高） - 控制流简单（少分支）
不适合 GPU 的程序： - 串行任务 - 大量分支判断 - 需要复杂同步
Q3 ： CUDA 和 OpenCL 有什么区别？
A ： - CUDA： NVIDIA 专有，功能丰富，生态完善 - OpenCL：开放标准，跨平台（ NVIDIA/AMD/Intel ），但功能相对简单
Q4 ：如何选择 Block 大小？
A ：一般原则： - 使用 32 的倍数（ Warp 大小） - 常见选择： 128, 256, 512 - 考虑寄存器使用量（太多线程会导致寄存器溢出）
📚 扩展阅读¶
- 《 CUDA 并行程序设计》 - 相关书籍
- NVIDIA CUDA 文档： docs.nvidia.com/cuda/
- 《 Programming Massively Parallel Processors 》 - David Kirk
🎯 下一步¶
学习 02-GPU 编程模型，深入理解 CUDA 线程层次和编程模型。