2、Hello World：第一个CUDA程序、主机与设备的概念、nvcc编译器、CUDA程序的基本结构

好，咱们正式开始写代码了。

说实话，我教过不少新人学CUDA。很多人一上来就被各种线程、块、网格搞晕了。我的建议是——先别管那些。咱们先跑通一个Hello World，感受一下GPU编程到底是怎么回事。

2.1 主机与设备：CPU和GPU的分工

在CUDA的世界里，有两个角色：主机（Host）和设备（Device）。

• 主机：就是CPU，负责控制逻辑、串行任务、数据搬运的调度。
• 设备：就是GPU，专门干并行计算的脏活累活。

你想想看，CPU像是个项目经理，GPU就是一群搬砖工。项目经理不亲自搬砖，但他得告诉工人们搬什么、搬到哪里。

我在项目中遇到过一个问题：有个同事把数据指针直接传给了GPU核函数，结果程序崩溃了。为什么？因为主机和设备的内存是物理隔离的，你不能直接访问对方的地址。嗯，这个坑我后面会详细讲。

2.2 第一个CUDA程序：Hello World

来，直接上代码。这是最经典的CUDA Hello World：

#include <stdio.h>

// 这是一个在GPU上运行的函数
__global__ void helloFromGPU() {
    printf("Hello World from GPU!\n");
}

int main() {
    // 主机端打印
    printf("Hello World from CPU!\n");

    // 调用GPU核函数
    helloFromGPU<<<1, 10>>>();

    // 等待GPU执行完毕
    cudaDeviceSynchronize();

    return 0;
}

这段代码很简单，但有几个关键点我得强调一下：

1. __global__ 修饰符：告诉编译器，这个函数要在GPU上跑，而且可以从CPU调用。
2. <<<1, 10>>>：这是CUDA的核函数调用语法。第一个参数是网格大小，第二个是块大小。这里我启动了10个线程。
3. cudaDeviceSynchronize()：让CPU等GPU干完活。没有这行，你可能看不到GPU的打印输出。

2.3 nvcc编译器：CUDA的幕后英雄

写好了代码，怎么编译？用 nvcc。

nvcc是NVIDIA提供的CUDA编译器。它会把 .cu 文件里的代码分成两部分：

• 主机代码（CPU部分）→ 交给gcc/clang编译
• 设备代码（GPU部分）→ 编译成PTX或cubin

编译命令很简单：

nvcc hello.cu -o hello

然后运行：

./hello

输出结果应该是：

Hello World from CPU!
Hello World from GPU!
Hello World from GPU!
...（一共10行）

为什么会打印10次？因为 <<<1, 10>>> 启动了10个线程，每个线程都执行了一次 printf。

2.4 CUDA程序的基本结构

一个标准的CUDA程序，通常包含这几个部分：

步骤	说明	示例代码
1. 分配主机内存	在CPU上申请空间	float *h_data = (float*)malloc(N * sizeof(float));
2. 分配设备内存	在GPU上申请空间	cudaMalloc(&d_data, N * sizeof(float));
3. 数据从主机拷贝到设备	把数据搬到GPU	cudaMemcpy(d_data, h_data, N * sizeof(float), cudaMemcpyHostToDevice);
4. 调用核函数	在GPU上执行计算	myKernel<<<grid, block>>>(d_data);
5. 数据从设备拷贝回主机	把结果拿回来	cudaMemcpy(h_data, d_data, N * sizeof(float), cudaMemcpyDeviceToHost);
6. 释放内存	清理资源	free(h_data); cudaFree(d_data);

这个流程，说白了就是：准备数据 → 送到GPU → 并行计算 → 取回结果 → 清理战场。我做了这么多年CUDA优化，90%的程序都逃不出这个框架。

2.5 避坑指南：新手最容易犯的错

嗯，这里我要多说几句。以下是我见过最多的错误：

• 忘记检查错误码：CUDA的API调用都会返回错误码。我建议每个API调用后都加 cudaGetLastError() 检查一下。
• 核函数里用 printf 太多：GPU上的 printf 很慢，调试时用用就好，正式代码里别留。
• 忘记同步：核函数是异步的，不调用 cudaDeviceSynchronize() 就访问结果，拿到的可能是垃圾数据。

2.6 小结

这一章我们干了三件事：

• 理解了主机和设备的分工
• 写了一个真正的CUDA程序
• 掌握了nvcc编译器的用法

下一章，我们会深入线程模型，看看那 <<<1, 10>>> 到底是怎么工作的。准备好了吗？