CUDA C++是C++的扩展，核心在于主机与设备协同：需装CUDA Toolkit和NVIDIA显卡，通过cudaMalloc/cudaMemcpy管理内存，用__global__定义kernel，按Grid-Block-Thread模型组织线程，并注意错误检查与同步。
用C++做GPU并行计算，核心是CUDA C++——它不是独立语言，而是C++的扩展，让你在标准C++代码里嵌入GPU核函数（kernel）和内存管理指令。关键不在于重学语法，而在于理解“主机（CPU）”和“设备（GPU）”的分工与协同。
一、环境准备：装好CUDA Toolkit和兼容显卡
CUDA只支持NVIDIA GPU（如GTX 10系以上、RTX系列、Tesla/A100等），且需安装对应版本的CUDA Toolkit（含编译器nvcc、运行时库、头文件）。Visual Studio（Windows）或GCC（Linux）作为主机编译器，nvcc负责把kernel部分编译成GPU可执行码，再链接进主程序。
- 确认显卡支持：终端运行 nvidia-smi 查驱动和GPU状态
- 下载CUDA Toolkit：匹配系统+驱动版本（官网提供详细兼容表）
- 设置PATH和LD_LIBRARY_PATH（Linux）或CUDA_PATH（Windows）
- 验证：编译运行 deviceQuery 示例，看到"Result = PASS"即成功
二、写第一个CUDA C++程序：向量加法
这是CUDA的“Hello World”。重点不是算法，而是结构范式：数据先从CPU内存拷贝到GPU显存 → 启动kernel在GPU上百线程并发执行 → 结果拷贝回CPU。
示例片段（简化版）：
立即学习“C++免费学习笔记（深入）”；
#include <cuda_runtime.h>
#include <iostream><p><strong>global</strong> void addVectors(float<em> a, float</em> b, float<em> c, int n) {
int idx = blockIdx.x </em> blockDim.x + threadIdx.x;
if (idx < n) c[idx] = a[idx] + b[idx];
}</p><p>int main() {
const int N = 1 << 20;
size_t size = N * sizeof(float);</p><pre class='brush:php;toolbar:false;'>// 1. 主机内存分配
float *h_a = new float[N], *h_b = new float[N], *h_c = new float[N];
// 2. 设备内存分配
float *d_a, *d_b, *d_c;
cudaMalloc(&d_a, size); cudaMalloc(&d_b, size); cudaMalloc(&d_c, size);
// 3. 数据拷贝到GPU
cudaMemcpy(d_a, h_a, size, cudaMemcpyHostToDevice);
cudaMemcpy(d_b, h_b, size, cudaMemcpyHostToDevice);
// 4. 定义执行配置：1024线程/块，足够覆盖N
int blockSize = 1024;
int gridSize = (N + blockSize - 1) / blockSize;
// 5. 启动kernel
addVectors<<<gridSize, blockSize>>>(d_a, d_b, d_c, N);
// 6. 同步等待完成（调试时必加）
cudaDeviceSynchronize();
// 7. 拷贝结果回CPU
cudaMemcpy(h_c, d_c, size, cudaMemcpyDeviceToHost);
// 清理...
delete[] h_a; delete[] h_b; delete[] h_c;
cudaFree(d_a); cudaFree(d_b); cudaFree(d_c);}
注意：__global__ 标记kernel函数；>> 是执行配置语法；cudaMemcpy方向必须明确（HostToDevice / DeviceToHost）；cudaDeviceSynchronize()确保kernel执行完再继续，避免读未写数据。
三、掌握线程组织模型：Grid-Block-Thread三层结构
CUDA把线程组织成三维层次：一个Grid（网格）包含多个Block（线程块），每个Block内有多个Thread（线程）。你通过 >> 控制启动多少Block和每个Block多少线程。
- threadIdx：线程在Block内的索引（x/y/z）
- blockIdx：Block在Grid内的索引
- blockDim：Block的维度大小（如dim3(16,16)表示16×16线程）
- gridDim：Grid的维度大小
- 常用计算全局索引：int idx = blockIdx.x * blockDim.x + threadIdx.x;（一维情形）
合理设置blockSize很重要：太小浪费资源，太大可能超SM（Streaming Multiprocessor）寄存器上限。常见取值为128、256、512、1024（必须是32的倍数，因Warp=32线程调度单位）。
四、避坑提醒：内存与同步常见错误
新手90%问题出在这两块：
- 忘记检查CUDA调用返回值：每条cudaMalloc/cudaMemcpy/kernel调用后加 cudaError_t err = cudaGetLastError(); if (err != cudaSuccess) printf("%s\n", cudaGetErrorString(err));
- 误用主机指针在kernel里：kernel中只能访问GPU内存（d_x），不能直接用h_x；所有输入输出都要显式拷贝
- 越界访问：kernel里务必加 if (idx ，因为gridSize常向上取整，最后Block可能有冗余线程
- 异步执行误解：kernel启动后CPU立刻往下走，不等它结束——需要cudaDeviceSynchronize()或cudaStreamSynchronize()显式同步
- 统一内存（Unified Memory）慎用：cudaMallocManaged简化了拷贝，但首次访问会触发迁移，对性能敏感场景建议手动管理
基本上就这些。CUDA C++不是魔法，它是把C++逻辑拆解成“CPU指挥 + GPU干活”的协作流程。写熟向量加法、矩阵乘、规约求和这几个典型模式，再结合Nsight Compute调试器看Occupancy、Memory Throughput指标，你就真正踏入GPU高性能计算的大门了。
