🖥️ GPU 计算与 CUDA 编程基础
为什么运维要学 CUDA?
云原生运维管理 GPU 工作负载时,理解 GPU 架构和 CUDA 编程模型有助于:
- 精准分配 GPU 资源(避免超配/浪费)
- 排查 GPU 工作负载性能问题
- 理解 GPU 共享/分片方案的底层原理
- 更好地与开发团队沟通 GPU 需求
本文整理自 CUDA PROGRAMMING(未完成)(NVIDIA CUDA Programming Guide 核心摘录)。
第一章:GPU 架构基础 🏗️
1.1 CPU vs GPU:设计哲学差异
| 维度 | CPU | GPU |
|---|---|---|
| 设计目标 | 快速执行单线程 | 并行执行数千线程 |
| 晶体管分配 | 数据缓存 + 流控 | 数据处理单元 |
| 并行度 | 数十个线程 | 数千个线程 |
| 单线程性能 | 高 | 较低(以量补质) |
| 适用场景 | 串行任务、低延迟 | 大规模并行计算 |
flowchart LR
subgraph "CPU"
A1["控制单元<br/>+ 缓存<br/>= 大部分芯片面积"]
A2["运算单元<br/>= 较小面积"]
end
subgraph "GPU"
B1["控制单元<br/>= 较小面积"]
B2["运算单元 × N<br/>= 大部分芯片面积"]
end
1.2 GPU 核心硬件层次
| 层次 | 说明 | 数量级 |
|---|---|---|
| GPC(图形处理簇) | SM 的集合 | 数个 |
| SM(流式多处理器) | 包含多个 CUDA 核心,实际执行单位 | 数十~数百 |
| Warp(线程束) | 32 个线程为一组,以锁步方式执行指令 | — |
| Thread(线程) | 最小的执行单元 | 数百万 |
Warp 是理解 GPU 性能的关键 GPU 以 32 线程/束(Warp) 为基本调度单位。Warp 内所有线程执行同一条指令→ 如果有分支分化(warp divergence),部分线程会被屏蔽,导致利用率下降。
1.3 SM 内部结构
| 组件 | 作用 |
|---|---|
| CUDA Cores | 整数/浮点运算单元 |
| Shared Memory | 片上可编程内存,极低延迟 |
| Registers | 寄存器文件,线程私有 |
| Warp Scheduler | 负责 warp 的调度与发射 |
| Load/Store Units | 内存访问单元 |
| Special Function Units | 特殊函数计算(sin/cos 等) |
第二章:CUDA 编程模型 🧠
2.1 线程层次结构
flowchart TD
subgraph "Grid(网格)"
direction LR
B1["Thread Block 0<br/>(线程块)"]
B2["Thread Block 1"]
B3["Thread Block 2"]
B4["..."]
end
subgraph "Thread Block(线程块)"
direction LR
T1["Thread 0"]
T2["Thread 1"]
T3["Thread 2"]
T4["..."]
end
Grid -->|每个 Block 在<br/>同一个 SM 执行| B1
B1 -->|Block 内的线程<br/>通过共享内存通信| T1
| 层次 | 说明 | 约束 |
|---|---|---|
| Grid | 一次内核启动的所有线程 | 无上限 |
| Thread Block | 一组线程,在同一个 SM 上执行 | 最多 1024 线程 |
| Thread | 最小的执行单位 | — |
2.2 内核(Kernel)的定义与启动
Kernel 是在 GPU 上执行的函数,使用 __global__ 声明:
// Kernel 定义:每个线程执行的代码
__global__ void vecAdd(float* A, float* B, float* C)
{
// 计算当前线程负责的元素索引
int workIndex = threadIdx.x + blockDim.x * blockIdx.x;
C[workIndex] = A[workIndex] + B[workIndex];
}
// 启动 Kernel
int main()
{
dim3 grid(4, 1, 1); // 4 个线程块
dim3 block(256, 1, 1); // 每块 256 线程
vecAdd<<<grid, block>>>(A, B, C);
}
| 内置变量 | 含义 | 示例 |
|---|---|---|
threadIdx | 线程在其块内的索引 | threadIdx.x |
blockIdx | 块在网格内的索引 | blockIdx.x |
blockDim | 线程块的维度 | blockDim.x = 256 |
gridDim | 网格的维度 | gridDim.x = 4 |
2.3 执行配置语法
// 语法:<<<网格维度, 块维度>>>
kernel<<<numBlocks, threadsPerBlock>>>(args);
// 1D 示例:1024 元素,每块 256 线程 → 4 块
vecAdd<<<4, 256>>>(A, B, C);
// 2D 示例:16×16 网格,8×8 块
dim3 grid(16, 16);
dim3 block(8, 8);
MatAdd<<<grid, block>>>(A, B, C);
第三章:GPU 内存层次 💾
3.1 内存全景
flowchart TD
subgraph "GPU 内存层次"
REG["寄存器<br/>片上 · 每个线程私有<br/>⚡ 最快"]
SMEM["共享内存<br/>片上 · 同一 Block 共享<br/>⚡ 快"]
GMEM["全局内存<br/>板载 DRAM · 所有线程可见<br/>🐢 较慢(但容量大)"]
LMEM["本地内存<br/>寄存器溢出时使用<br/>📍 实际存储在全局内存"]
end
REG -->|容量不足时| LMEM
SMEM -->|手动管理| GMEM
| 内存类型 | 位置 | 作用域 | 延迟 | 容量 |
|---|---|---|---|---|
| 寄存器 | 片上 | 线程私有 | 极低 | 非常有限 |
| 共享内存 | 片上 | 同一 Block 内所有线程 | 极低 | 几十 KB |
| 全局内存 | 板载 DRAM | 所有线程 + Host | 较高 | 数 GB~数十 GB |
| 本地内存 | 板载 DRAM | 线程私有 | 较高 | 有限 |
| 常量内存 | 板载 DRAM(带缓存) | 所有线程只读 | 中 | 64 KB |
| 纹理内存 | 板载 DRAM(带缓存) | 所有线程只读 | 中 | 有限 |
3.2 统一内存(Unified Memory)
CUDA 6.0+ 引入,CPU 和 GPU 共享同一地址空间,无需手动拷贝数据。
// 传统方式:显式内存管理
cudaMalloc(&devA, size); // GPU 分配
cudaMemcpy(devA, hostA, size, ...); // CPU→GPU 拷贝
kernel<<<...>>>(devA);
cudaMemcpy(hostA, devA, size, ...); // GPU→CPU 拷贝
// 统一内存方式
cudaMallocManaged(&A, size); // CPU/GPU 自动迁移
kernel<<<...>>>(A);
cudaDeviceSynchronize(); // 等待 GPU 完成
// A 的数据自动对 CPU 可见
3.3 关键性能概念:合并访问(Coalescing)
当 Warp 中 32 个线程访问连续内存地址时,GPU 会将它们合并为最少的内存事务。
非连续访问 → 多次内存事务 → 带宽浪费。
第四章:SIMT 执行模型 ⚡
4.1 Warp 执行
| 概念 | 说明 |
|---|---|
| Warp 大小 | 固定 32 线程 |
| 锁步执行 | Warp 内所有线程执行同一条指令 |
| Warp Divergence | 线程走向不同分支 → 部分被屏蔽 → 利用率下降 |
| 优化建议 | Block 大小最好是 32 的倍数 |
4.2 线程块调度
flowchart LR
subgraph "GPU SM"
SM1["SM 1<br/>Block 0<br/>Block 1<br/>Block 2"]
SM2["SM 2<br/>Block 3<br/>Block 4<br/>Block 5"]
end
subgraph "Grid"
B0["Block 0"] --> SM1
B3["Block 3"] --> SM2
B1["Block 1"] --> SM1
B4["Block 4"] --> SM2
end
Block 在 SM 之间的分配顺序没有保证,Block 之间不能相互依赖。
4.3 线程块簇(Thread Block Clusters,计算能力 9.0+)
| 特性 | 说明 |
|---|---|
| 引入版本 | 计算能力 9.0(Hopper 架构) |
| 用途 | 同一簇内的 Block 可在同一 GPC 上协同调度 |
| 最大簇大小 | 8 个 Block(可移植) |
| 通信 | 支持跨 Block 同步 + 分布式共享内存 |
第五章:错误检查与调试 🔍
5.1 CUDA 错误检查
每个 CUDA API 返回 cudaError_t 类型:
#define CUDA_CHECK(expr) do { \
cudaError_t result = expr; \
if (result != cudaSuccess) { \
fprintf(stderr, "CUDA Error: %s:%i: %s\n", \
__FILE__, __LINE__, cudaGetErrorString(result)); \
} \
} while(0)
// 使用示例
CUDA_CHECK(cudaMalloc(&devA, size));
kernel<<<grid, block>>>(devA);
CUDA_CHECK(cudaGetLastError()); // 检查启动参数错误
CUDA_CHECK(cudaDeviceSynchronize()); // 检查执行错误
5.2 异步错误陷阱
| 类型 | 说明 | 检查方式 |
|---|---|---|
| 同步错误 | 参数错误(如 Block 过大) | 启动后立即 cudaGetLastError |
| 异步错误 | 执行时错误(越界访问) | cudaDeviceSynchronize 后检查 |
5.3 CUDA_LOG_FILE 环境变量
# 设置后,CUDA 驱动将错误信息写入指定文件
$ env CUDA_LOG_FILE=cudaLog.txt ./my_cuda_app
$ cat cudaLog.txt
# → 包含更详细的错误信息
第六章:K8s 运维视角 🎯
6.1 CUDA 知识如何辅助 GPU 调度
| 运维场景 | 需要的 CUDA 知识 |
|---|---|
| GPU 资源分配 | 理解显存需求 → 合理设置 nvidia.com/gpu |
| GPU 共享 | 理解 SM 与 Block 的关系 → MPS/MIG 分割策略 |
| 性能诊断 | 理解 Warp、合并访问 → 判断 workload 是否适配 GPU |
| 驱动版本选择 | 不同计算能力(Compute Capability)对应的驱动要求 |
| 故障排查 | CUDA 错误码含义、内存越界等 |
6.2 关键术语对照
| CUDA 术语 | 运维理解 |
|---|---|
| Compute Capability | GPU 硬件版本号,决定支持的 CUDA 特性 |
| SM | GPU 的实际计算单元,类比 CPU 核心 |
| 显存/VRAM | GPU 上的全局内存,决定了可加载的模型大小 |
| Warp | 最小调度单位(32 线程),理解它是优化性能的关键 |
| Kernel | GPU 上执行的函数,对应 Pod 中的 GPU workload |
| cudaMallocManaged | 统一内存,简化 GPU 编程但可能影响性能 |
📚 深入学习路径
- 阅读原始文档 → CUDA PROGRAMMING(未完成)(94KB 完整摘录)
- 动手实践 → NVIDIA 官方 CUDA 示例代码
- 调度实践 → K8s 中
nvidia-device-plugin+ GPU 共享方案 - 进阶 → NVIDIA MIG(多实例 GPU)、MPS(多进程服务)
关联文档
- CUDA PROGRAMMING(未完成) — 原始完整摘录
- GPU 调度与 cgroup 问题(文档待创建) — K8s GPU 调度问题
- ../kubernetes/core/K8s 核心概念 — K8s 基础
- ../programming/脚本与编程 — Python/SDK 调用 CUDA
最后更新:2026-06-02