ばぐだし

続いてエクササイズ２、カーネルの起動いきまーす。

ソース一部抜粋

// Part 3 of 5: implement the kernel
__global__ void myFirstKernel(int *d_a)
{
    int idx = blockIdx.x*blockDim.x + threadIdx.x;
    d_a[idx] = 1000*blockIdx.x + threadIdx.x;
}

    // Part 2 of 5: configure and launch kernel
    dim3 dimGrid(numBlocks);
    dim3 dimBlock(numThreadsPerBlock);
    myFirstKernel<<<dimGrid,dimBlock>>>(d_a);

    // block until the device has completed
    cudaThreadSynchronize();

※以下私なりに丸めたちょっと怪しい表現を使います。正しい表現は本家ページで！
CUDAでは「関数名<<<グリッド内のブロックの数、１ブロックあたりのスレッドの数>>>(引数) 」で並列カーネル(デバイス上で実行される機能)を呼び出す。
CUDAで書いた関数(この場合myFirstKernel)が1つのスレッドに当たる。1度に実行できるのは1つのグリッドだけ。従って、一度に実行されるスレッド数(myFirstKernel)は、グリッド内のブロックの数(numBlocks)×１ブロックあたりのスレッドの数(numThreadsPerBlock)となる。
ちなみに、グリッド内のブロックの数、１ブロックあたりのスレッドの数は、2次元での表現も可能(ブロックあたりスレッド数は3次元まで可能)で、画像屋さんにはとても助かる。

また、カーネル内では、以下の変数が自動的に定義されており、アクセスができる。
dim3 gridDim;
ブロックのグリッドの次元（最大で2次元）
dim3 blockDim;
スレッドのブロックの次元
dim3 blockIdx;
グリッド内のブロックのインデックス
dim3 threadIdx;
ブロック内のスレッドのインデックス
てなわけでスレッドのインデックスとしては、

int idx = blockDim.x * blockIdx.x + threadIdx.x;

が1次元の並列カーネル呼び出しでは定番となる。

以上、次もエクササイズ！