グリッドを 1 次元から 2 次元グリッドに拡張しようとしています。これを行う方法はありますか?
これが私の現在のコードです:
int idx = threadIdx.x + blockDim.x * blockIdx.x;
リストには、#include
次の定義があります。
#define BLOCKS_PER_GRID 102
#define THREADS_PER_BLOCK 1024
グリッドを 1 次元から 2 次元グリッドに拡張しようとしています。これを行う方法はありますか?
これが私の現在のコードです:
int idx = threadIdx.x + blockDim.x * blockIdx.x;
リストには、#include
次の定義があります。
#define BLOCKS_PER_GRID 102
#define THREADS_PER_BLOCK 1024
1 次元のスレッドブロックの 2D グリッドを使用して、グリッドの次元ごとに 65535 ブロックという制限を回避することもできます (cc3.0 より前のデバイスの場合)。これは、データの 2 次元配列表現を導入せずに、基本的に 1 次元の問題を限界を超えて拡張する簡単な方法である可能性があります。
DATA_ELEMENTS
カーネルが処理する要素の数 (スレッドごとに 1 つの要素) として定義されたパラメーターがあるとします。DATA_ELEMENTS
が 65535*1024 より大きい場合、各スレッドが 1 つの要素しか処理しない場合、1 次元グリッドを使用してすべてを処理することはできません。
パラメータはそのままにしておくことができますTHREADS_PER_BLOCK
。カーネル内のスレッド インデックスの計算は、次のように変更されます。
int idx = threadIdx.x + (blockDim.x * ((gridDim.x * blockIdx.y) + blockIdx.x));
次のようなものを使用して、カーネル計算を確実に調整する必要があります。
if (idx < DATA_ELEMENTS){
(kernel code)
}
グリッドの寸法は次のようになります。
dim3 grid;
if (DATA_ELEMENTS > (65535*THREADS_PER_BLOCK)){ // create a 2-D grid
int gridx = 65535; // could choose another number here
int gridy = ((DATA_ELEMENTS+(THREADS_PER_BLOCK-1))/THREADS_PER_BLOCK)/gridx;
if ((((DATA_ELEMENTS+(THREADS_PER_BLOCK-1))/THREADS_PER_BLOCK)%gridx) != 0) gridy++;
grid.x=gridx;
grid.y=gridy;
grid.z=1;
}
else{ // create a 1-D grid
int gridx = (DATA_ELEMENTS+(THREADS_PER_BLOCK-1))/THREADS_PER_BLOCK;
grid.x=gridx;
grid.y=1;
grid.z=1;
}
カーネルを次のように起動します。
kernel<<<grid, THREADS_PER_BLOCK>>>(...);
この種の問題に取り組む別の方法は、ある次元の 1-D グリッドを作成し (グリッド内のスレッドの総数NUM_THREADS_PER_GRID
をfor ループや while ループのようなものを使用します。
while (idx < DATA_ELEMENTS) {
(code to process an element)
idx += NUM_THREADS_PER_GRID
}
ブロックごとに 1024 スレッドが必要な場合、ブロックは簡単に 2D に再形成できます。
32 x 32 = 1024;
したがって、ブロックは次のようになります。
dim3 Block(32,32); //1024 threads per block. Will only work for devices of at least 2.0 Compute Capability.
正確な要件はわかりませんが、通常、ブロックの数は固定されていません (マクロで定義したように)。ブロックの数は、グリッドが動的にスケーリングされるように、入力データのサイズによって異なります。
あなたの場合、多くのオプションがありますが、グリッドに最も近い最適なサイズは17 x 6
またはになり6 x 17
ます。
dim3 Grid(17,6);
これで、次のパラメーターを使用してカーネルを呼び出すことができます。
kernel<<<Grid,Block>>>();
カーネル内では、スレッドの 2 次元インデックスが次のように計算されます。
int xIndex = blockIdx.x * blockDim.x + threadIdx.x;
int yIndex = blockIdx.y * blockDim.y + threadIdx.y;
または、x/y の代わりに行/列の規則に従う場合は、次のようになります。
int row = blockIdx.y * blockDim.y + threadIdx.y;
int column = blockIdx.x * blockDim.x + threadIdx.x;