0

以下のプログラムで、行列の指標を下記のように書ける理由が分かりません。
CUDAでは行列はGPUにどのように割り当てられるのでしょうか?
thread一つ一つに行列の各要素が割り当てられるのでしょうか?

// Kernel definition
__global__ void MatAdd(float A[N][N], float B[N][N],
float C[N][N])
{
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    int j = blockIdx.y * blockDim.y + threadIdx.y;
    if (i < N && j < N)
        C[i][j] = A[i][j] + B[i][j];
}

int main()
{
    ...
    // Kernel invocation
    dim3 threadsPerBlock(16, 16);
    dim3 numBlocks(N / threadsPerBlock.x, N / threadsPerBlock.y);
    MatAdd<<<numBlocks, threadsPerBlock>>>(A, B, C);
    ...
}

1 件の回答 1

1

厳密なことを省いて概要だけを説明しますと、MatAdd<<<numBlocks, threadsPerBlock>>>(A, B, C);で渡されている、 numBlocks, threadsPerBlock分、スレッドが生成されます。この場合、N個のスレッドが生成されます。

N個のスレッドがそれぞれ、カーネルコードであるMatAdd()を実行します。つまり、 MatAdd()がN個のスレッドで個別に処理されています。

このカーネルには、スレッド固有の情報がCUDAによって付与されています。

blockDim.xはX軸のスレッドブロック数、 blockIdx.xは今実行されているスレッドのX軸のスレッドブロックのインデックス値、thread.xは今実行されているスレッドのインデックス値が入ります。Y軸も同様です。

各スレッドは上記の値を使ってどの配列にアクセスすべきかを決めています。
それが

int i = blockIdx.x * blockDim.x + threadIdx.x;
int j = blockIdx.y * blockDim.y + threadIdx.y;

の部分です。

ですので、配列サイズよりも大きいスレッドを生成した場合、簡単にメモリ領域外アクセスを起こします。このため配列サイズでのアクセスチェックは必ず行ってください。

スレッドはいくらでも呼び出すことができます(上限はありますが)し、実際にはWarpと呼ばれるスレッドグループで32スレッドずつスレッドが実行されますので注意が必要です。

1
  • 回答ありがとうございます!各スレッド毎にどの配列要素にアクセスするか、ブロックとスレッドそれぞれの指標で指定しているんですね。理解できました!
    – yutaiguhel
    2019年7月17日 9:38

この質問に回答するには、ログインする必要があります。

求めていた回答ではありませんか? のタグが付いた他の質問を参照する。