本章介绍了CUDA编程模型背后的主要概念,概述了它们在C++中是如何公开的。在编程接口中对CUDA C++进行了详细的描述。
本章和下一章中使用的矢量加法示例的完整代码可以在矢量加法CUDA示例中找到。
2.1、内核
CUDA C++通过允许程序员定义称为内核的C++函数来扩展C++,这些函数在被调用时由N个不同的CUDA线程并行执行N次,而不是像常规C++函数那样只执行一次。
内核是使用__global__声明说明符定义的,并且为给定内核调用执行该内核的CUDA线程数是使用新的<<…>>指定的执行配置语法(请参阅C++语言扩展)。执行内核的每个线程都有一个唯一的线程ID,可以通过内置变量在内核中访问该ID。如图所示,以下示例代码使用内置变量threadIdx,将大小为N的两个矢量A和B相加,并将结果存储到矢量C中:
// Kernel definition
__global__ void VecAdd(float* A, float* B, float* C)
{
int i = threadIdx.x;
C[i] = A[i] + B[i];
}
int main()
{
...
// Kernel invocation with N threads
VecAdd<<<1, N>>>(A, B, C);
...
}
这里,执行VecAdd()的N个线程中的每一个执行一对加法。
2.2、线程层次结构