GPU编程--kernels（2）

1. “如何区分不同的数据单位单位呢？”
2. “如何确定程序是在CPU端执行，还是GPU端执行呢？”
3. “如何确定要调用的GPU线程数呢？”

下面举一个真实场景理解一下上述问题：用GPU将一幅图像的每个像素分别减去均值。我们可以将每一个像素影射到一个GPU线程，所有线程并行完成减均值操作。问题来了，“线程”如何区分不同的像素呢？在c语言中又该如何表示这种操作呢？

先代码，后解释

// Kernel definition
__global__ void VecAdd(float* A, float mean, float* C)
{
int i = threadIdx.x;
C[i] = A[i] - mean;
}
int main()
{
...
// Kernel invocation with N threads
VecAdd<<<1, N>>>(A, mean, C);
...
}

1. 每一个线程都有一个唯一的thread ID，这可以通过内建变量threadIdx获取，因为我们将每一个像素都影射到了一个线程，这样线程号就可以作为不同像素之间的区分了
2. 在GPU端执行的函数，CUDA用__global__关键字标识。其本质就是对c语言进行了扩展，扩展的部分有明确的含义，就是在GPU端执行的函数
3. 在调用GPU函数时，可以在GPU函数后紧跟<<<...>>>，用来指定所需的线程数。

对于假设的场景，我们可以这样做（其参考代码）：用<<<1,N>>>线程数，其实也就是像素的个数，用线程ID threadIdx.x区分不同的像素，最后并行完成减均值操作。

在对CUDA有了一个感性理解之后，下面的问题是：如何通过概念的构造充分利用有限的GPU资源。