cudaで3つのC++ループを解く

Question

私は現在、大きな2D配列の2Dブロックを取り出し、小さな2D配列に配置するプログラムを作成しています。配列は実際には1D配列として保存されますが、それらはまっすぐに整列された2D配列であると仮定します。今のところ、ループのために大きな配列の新しい部分に移動する前に、大きなループのNx、Ny要素を取り出してください。

私は実際にこのデータ転送をCUDAカーネルでより効率的に行う方法を見つけられません。私は同じサイズの配列だけを持っていれば解決策を作ることができただろう。しかし、それぞれのNx * Ny要素の後に新しい場所があると、私は混乱します。

以下は3つのforループです。 posxとposyには、開始位置を減算する必要がある座標が含まれます。

for (int i = 0; i < loadsize; i++) 
    { 
     for (int k = 0; k < Searchsizey; k++) 
     { 
      for (int l = 0; l < Searchsizex; l++) 
      { 
       img[count] = ImgInt[posx[i] - ImgStartx + (posy[i] - ImgStarty) * sizex + sizex*k + l]; 
       count++; 
      } 
     } 
    } 

問題は、2つの内部ループが最大SearchsizexとSearchsizeyまで実行されることです。 lとkをスレッドからのインデックスとしてカーネルに実装すると、これらはSearchsizexとSearchsizeyよりも大きくなります。

私はトラフiを繰り返すCUDAカーネルを作ろうと考えましたが、このようにするのは非常に効率的です。以下は私がカーネルの中に入れていることについての話ですが、あまり好きではありません。その唯一の「セミパラレル」です。

あなたは、これを並列化する方法についていくつかの提案がありますように、ありがとう！

int l = blockIdx.x*blockDim.x + threadIdx.x; 
int k = blockIdx.y*blockDim.y + threadIdx.y; 
     for (int i = 0; i < loadsize; i++){ 
      img[l + k*sizex + (i*sizex*sizey)] = ImgInt[posx[i] - ImgStartx + (posy[i] - ImgStarty) * sizex + sizex*k + l];   
     } 

Answer 1

このスレッドに返信がありませんでしたので、私は答えとしてFlorentsのコメントを使用します。

書き込みはfloat4型である必要があり、ターゲットハードウェアによっては、ImgIntにテクスチャまたはサーフェスのデータ型を使用する必要があります。

それ以外は、並列性のアプローチは非常に賢明です。