2. CUDA编程基础：GPU架构、线程层次、内存层次、CUDA核函数编写

2.1 GPU架构核心认知：从“多核”到“众核”

CPU与GPU的设计哲学截然不同。CPU拥有少量但强大的核心（如8核16线程），擅长处理复杂分支和串行任务；GPU则集成了数千个轻量级计算核心，专为大规模并行、高吞吐量的数据密集型计算设计。理解这一差异是编写高效CUDA代码的前提。

现代NVIDIA GPU（如Ampere、Hopper架构）的硬件层次可概括为：

• 流式多处理器（SM）：GPU的基本计算单元，每个SM包含若干CUDA核心、共享内存、寄存器文件和调度器。
• CUDA核心：执行浮点/整数运算的最小单元，一个SM通常包含64~128个CUDA核心。
• 线程束（Warp）：SM以32个线程为一组进行调度和执行，这是硬件执行的基本粒度。
• 全局内存（Global Memory）：所有线程共享的大容量显存（如HBM），延迟高（数百周期），需通过合并访问提升带宽利用率。

2.2 线程层次与索引计算

CUDA将线程组织为三层结构：网格（Grid）→ 线程块（Block）→ 线程（Thread）。核函数启动时，需要指定网格和线程块的维度。每个线程通过内置变量确定自己的全局唯一ID，这是访问数据的基础。

// 核函数启动配置
dim3 gridDim(4, 2);      // 网格：4×2 = 8个线程块
dim3 blockDim(16, 16);   // 线程块：16×16 = 256个线程
kernel<<<gridDim, blockDim>>>(data, N);

// 线程全局ID计算（一维示例）
int tid = blockIdx.x * blockDim.x + threadIdx.x;
if (tid < N) {  // 边界检查，防止越界
    data[tid] = data[tid] * 2.0f;
}

2.3 内存层次与数据访问策略

CUDA内存模型是分层的，不同内存的速度、容量、作用域差异巨大。合理利用内存层次是性能优化的核心。

2.4 CUDA核函数编写规范与案例

核函数（Kernel）是运行在GPU上的函数，需遵循以下规则：

• 使用 __global__ 修饰符声明，返回类型必须为 void
• 不能使用可变参数（如 printf 受限）
• 不能调用主机端函数（如 malloc）
• 通过 <<<grid, block>>> 语法启动

下面是一个完整的向量加法案例，展示核函数编写、内存分配和数据传输流程：

// 核函数定义
__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];
    }
}

// 主机端调用
int main() {
    int n = 1 << 20;  // 1M 元素
    size_t bytes = n * sizeof(float);

    // 1. 分配主机内存并初始化
    float *h_A = (float*)malloc(bytes);
    float *h_B = (float*)malloc(bytes);
    float *h_C = (float*)malloc(bytes);
    // ... 初始化 h_A, h_B ...

    // 2. 分配设备内存
    float *d_A, *d_B, *d_C;
    cudaMalloc(&d_A, bytes);
    cudaMalloc(&d_B, bytes);
    cudaMalloc(&d_C, bytes);

    // 3. 主机→设备数据传输
    cudaMemcpy(d_A, h_A, bytes, cudaMemcpyHostToDevice);
    cudaMemcpy(d_B, h_B, bytes, cudaMemcpyHostToDevice);

    // 4. 启动核函数
    int threads = 256;
    int blocks = (n + threads - 1) / threads;
    vectorAdd<<<blocks, threads>>>(d_A, d_B, d_C, n);

    // 5. 设备→主机数据传输
    cudaMemcpy(h_C, d_C, bytes, cudaMemcpyDeviceToHost);

    // 6. 释放内存
    cudaFree(d_A); cudaFree(d_B); cudaFree(d_C);
    free(h_A); free(h_B); free(h_C);
    return 0;
}

2.5 性能误区与调优复盘