167

GPU によって実行されるスレッドはどのように編成されますか?

4

2 に答える 2

302

ハードウェア

たとえば、GPU デバイスに 4 つのマルチプロセッシング ユニットがあり、それぞれが 768 のスレッドを実行できる場合: 特定の時点で、実際に並行して実行されるスレッドは 4*768 以下です (より多くのスレッドを計画した場合、それらは待機します)。彼らの番）。

ソフトウェア

スレッドはブロックで編成されます。ブロックは、マルチプロセッシング ユニットによって実行されます。ブロックのスレッドは、1Dimension(x)、2Dimensions (x,y)、または 3Dim インデックス (x,y,z) を使用して識別 (インデックス付け) できますが、いずれの場合も、この例では x y z <= 768 (他の制限が適用されます) x、y、z については、ガイドとデバイスの機能を参照してください)。

明らかに、これらの 4*768 スレッドよりも多くのスレッドが必要な場合は、4 ブロックより多くが必要です。ブロックは、1D、2D、または 3D でインデックス付けすることもできます。GPU に入るのを待っているブロックのキューがあります (この例では、GPU に 4 つのマルチプロセッサがあり、4 つのブロックのみが同時に実行されているため)。

簡単なケース: 512x512 画像の処理

1 つのスレッドで 1 つのピクセル (i,j) を処理するとします。

それぞれ 64 スレッドのブロックを使用できます。次に、512*512/64 = 4096 ブロックが必要です (つまり、512x512 スレッド = 4096*64 になります)

blockDim = 8 x 8 (ブロックあたり 64 スレッド) を持つ 2D ブロック内のスレッドを整理する (イメージのインデックス作成を容易にする) のが一般的です。私はそれをthreadsPerBlockと呼んでいます。

dim3 threadsPerBlock(8, 8);  // 64 threads

および 2D gridDim = 64 x 64 ブロック (4096 ブロックが必要)。私はこれを numBlocks と呼んでいます。

dim3 numBlocks(imageWidth/threadsPerBlock.x,  /* for instance 512/8 = 64*/
              imageHeight/threadsPerBlock.y);

カーネルは次のように起動されます。

myKernel <<<numBlocks,threadsPerBlock>>>( /* params for the kernel function */ );

最後に、「4096 ブロックのキュー」のようなものがあり、GPU のマルチプロセッサの 1 つにブロックが割り当てられて 64 のスレッドが実行されるのを待っています。

カーネルでは、スレッドによって処理されるピクセル (i,j) は次のように計算されます。

uint i = (blockIdx.x * blockDim.x) + threadIdx.x;
uint j = (blockIdx.y * blockDim.y) + threadIdx.y;
于 2010-03-06T11:16:49.243 に答える
10

9800GT GPU を想定します。

  • 14 個のマルチプロセッサ (SM) を搭載
  • 各 SM には 8 つのスレッド プロセッサ (AKA ストリーム プロセッサ、SP またはコア) があります。
  • ブロックごとに最大 512 のスレッドを許可
  • warpsize は 32 です (つまり、14x8=112 個のスレッド プロセッサのそれぞれが最大 32 個のスレッドをスケジュールできます)。

https://www.tutorialspoint.com/cuda/cuda_threads.htm

ブロックは 512 を超えるアクティブなスレッドを持つことができないため、__syncthreads限られた数のスレッドしか同期できません。つまり、600 スレッドで以下を実行した場合:

func1();
__syncthreads();
func2();
__syncthreads();

その場合、カーネルは 2 回実行する必要があり、実行順序は次のようになります。

  1. func1 は最初の 512 スレッドで実行されます
  2. func2 は最初の 512 スレッドで実行されます
  3. func1 は残りのスレッドに対して実行されます
  4. func2 は残りのスレッドに対して実行されます

ノート：

要点は__syncthreads、ブロック全体の操作であり、すべてのスレッドを同期するわけではありません。

__syncthreads512 を超えるスレッドでブロックを作成し、ワープにスケジューリングを処理させることができるため、同期できるスレッドの正確な数についてはわかりません。私の理解では、func1 は少なくとも最初の 512 スレッドで実行されると言った方が正確です。

この回答を編集する前 (2010 年に戻る)、14x8x32 のスレッドが を使用して同期されていることを測定しまし__syncthreadsた。

より正確な情報を得るために、誰かがこれをもう一度テストしていただければ幸いです。

于 2010-06-14T06:25:52.623 に答える