CUDAカーネルのさまざまなスレッドグループ

Question

CUDAカーネルを使用して、Cで相互相関関数を高速化しようとしています。今のところ、これは私が持っているものです。

__global__ void xcorr(cuDoubleComplex *temp1, cuDoubleComplex *temp2, cuDoubleComplex *temp3, int Nb, int binM, int Nspb) 
{ 
    for (int k1 = 0; k1 < Nb; k1++) 
    { 
     int idx = blockIdx.x * blockDim.x + threadIdx.x; 
     for (int j1 = 0; j1 < Nspb; j1++) 
     { 
      if ((j1 + idx) <(Nspb + binM)) 
      { 
       temp3[idx + k1*(binM + 1)].x += (temp1[idx + j1 + (k1*(binM + Nspb))].x*temp2[j1 + (k1*Nspb)].x) + (temp1[idx + j1 + (k1*(binM + Nspb))].y*temp2[j1 + (k1*Nspb)].y); 
       temp3[idx + k1*(binM + 1)].y += (-temp1[idx + j1 + (k1*(binM + Nspb))].x*temp2[j1 + (k1*Nspb)].y) + (temp1[idx + j1 + (k1*(binM + Nspb))].y*temp2[j1 + (k1*Nspb)].x); 
      } 
     } 
    } 
} 

結果は、私が期待したものですが、それはまだ50 secondsの周りに、動作させるために時間がかかっ。私はカーネルを召喚するとき、私はそれを行うこのよう

xcorr << <1, 1000 >> > (cuda_E2, cuda_A2, cuda_temp, Nb, *binM, Nspb);

そして、何私が考えていたことはj1（Nspb=5000）でループを回避するために、1000件のスレッドごとに、代わりに一つだけの6つのブロックを送信することです。私はさまざまな方法で試しましたが、2つの異なるスレッドグループを使用する方法を見つけることができません。最初のブロックは同じ方法で、もう1つはj1ループを置き換えます。誰かが私にどのように表示できますか？

ご協力いただければ幸いです。

Answer 1

最初のブロックと他のブロックの異なるコードを書き、if（blockIdx.x == 0）の枝に入れ、< < < 6,1000 >>>グリッドを開始しますか？

__global__ void xcorr(...) 
{ 
    if (blockIdx.x==0) { 
     // do block zero stuff 
    } 
    else { 
     // what the other blocks shall do 
    } 
} 

1Dのconvul​​ution /相互相関のために一般的には

、：

1. がすべてのスレッドブロックが同じことをやらせる一定のメモリ
2. にカーネルを置くが、結果の配列の異なるタイルの
3. 各タイルに、このタイルに必要な入力配列のすべての要素を共有メモリにロードさせます（エッジの「ハロー」を含む）。