gpu-computing

GPU 计算与 CUDA 编程基础

·14 分钟阅读·5366 字
cudagpunvidiaparallel-computingcloud-nativereference

面向云原生运维工程师的 GPU 计算入门：GPU 架构、CUDA 编程模型、内存层次与 K8s GPU 调度基础

📋 目录

🖥️ GPU 计算与 CUDA 编程基础

为什么运维要学 CUDA？

云原生运维管理 GPU 工作负载时，理解 GPU 架构和 CUDA 编程模型有助于：

  • 精准分配 GPU 资源（避免超配/浪费）
  • 排查 GPU 工作负载性能问题
  • 理解 GPU 共享/分片方案的底层原理
  • 更好地与开发团队沟通 GPU 需求

本文整理自 CUDA PROGRAMMING（未完成）（NVIDIA CUDA Programming Guide 核心摘录）。

第一章：GPU 架构基础 🏗️

1.1 CPU vs GPU：设计哲学差异

维度CPUGPU
设计目标快速执行单线程并行执行数千线程
晶体管分配数据缓存 + 流控数据处理单元
并行度数十个线程数千个线程
单线程性能高较低（以量补质）
适用场景串行任务、低延迟大规模并行计算
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 CapabilityGPU 硬件版本号，决定支持的 CUDA 特性
SMGPU 的实际计算单元，类比 CPU 核心
显存/VRAMGPU 上的全局内存，决定了可加载的模型大小
Warp最小调度单位（32 线程），理解它是优化性能的关键
KernelGPU 上执行的函数，对应 Pod 中的 GPU workload
cudaMallocManaged统一内存，简化 GPU 编程但可能影响性能

📚 深入学习路径

  1. 阅读原始文档 → CUDA PROGRAMMING（未完成）（94KB 完整摘录）
  2. 动手实践 → NVIDIA 官方 CUDA 示例代码
  3. 调度实践 → K8s 中 nvidia-device-plugin + GPU 共享方案
  4. 进阶 → NVIDIA MIG（多实例 GPU）、MPS（多进程服务）

关联文档

  • CUDA PROGRAMMING（未完成） — 原始完整摘录
  • GPU 调度与 cgroup 问题（文档待创建） — K8s GPU 调度问题
  • ../kubernetes/core/K8s 核心概念 — K8s 基础
  • ../programming/脚本与编程 — Python/SDK 调用 CUDA

最后更新：2026-06-02

返回首页订阅更新