opencl - opencl 同期

Question

私はopenclが初めてで、バリア機能について理解できないことがあるようです。これは私のカーネルのコードです。これは、出力が *w の標準的な行列ベクトル計算です。ベクトルの次元と同じ 64 個の作業単位を持つ 1 つの作業グループがあります。

#pragma OPENCL EXTENSION cl_khr_fp64 : enable
__kernel void fmin_stuff(__global double *h, __global double *g, __global double  
  *w,int n,__global int * gid) {

// Get the index of the current element
int i = get_global_id(0);
int j;
gid[i]=get_local_id(0);

w[i]=-g[i];
barrier(CLK_GLOBAL_MEM_FENCE | CLK_LOCAL_MEM_FENCE);
for (j=0;j<n;j++)
{
  if (j<i)
    w[i]-=h[i+j*n]*w[j];
  barrier(CLK_GLOBAL_MEM_FENCE | CLK_LOCAL_MEM_FENCE);
}
}

問題は、コードがランダムに失敗することです。出力はしばらくの間正しいです。各実行の w の初期値は次のとおりです。

-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.34999 2.51524 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.10141 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.72261 2.80155 
-0.148351 -0.309007 0.133204 -1.39589 2.88335 -2.68636 2.77369

プログラムは、いずれの場合もカーネルが正常に実行されたことを報告します。すべての実行で、ベクトル w の値は最終的に不正確になります。アドバイスをいただければ幸いです。

これが単純な行列の乗算であるかどうかについて、いくつかの混乱がありました。そうではない。これは、w の最初の 5 つの項のみを含めるコードで達成しようとしていることです。

w(1)=-g(1);
w(2)=-g(2);
w(3)=-g(3);
w(4)=-g(4);
w(5)=-g(5);

w(2)-=h(2)*w(1);
w(3)-=h(3)*w(1);
w(4)-=h(4)*w(1);
w(5)-=h(5)*w(1);

w(3)-=h(3+N)*w(2);
w(4)-=h(4+N)*w(2);
w(5)-=h(5+N)*w(2);

w(4)-=h(4+2*N)*w(3);
w(5)-=h(5+2*N)*w(3);

w(5)-=h(5+3*N)*w(4);

また、カーネルは、プログラムの実行ごとに 1 回だけ呼び出されます。ランダムな動作は、プログラムを複数回実行した結果です。

コメントのおかげで、自分が何を間違っていたのかがわかりました。ワークグループとアイテムを次のように構成しました

size_t global_item_size[3] = {N, 1, 1}; // Process the entire lists
size_t local_item_size[3] = {1,1,1}; // Process in groups of 64
ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL,
        global_item_size, local_item_size, 0, NULL, NULL);

あるべきだったとき。

size_t global_item_size[3] = {N, 1, 1}; // Process the entire lists
size_t local_item_size[3] = {N,1,1}; // Process in groups of 64
ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL,
        global_item_size, local_item_size, 0, NULL, NULL);

助けてくれてありがとう。これは私にとっては素晴らしいことですが、おそらく他の人にとってはあまり興味がありません。

score 0 · Accepted Answer

まず第一に、あなたの場合、CLK_LOCAL_MEM_FENCEを使用する必要はありません。

しかし、私はコピーすることをお勧めします

グローバル->ローカル
ローカルデータを操作する
ローカル->グローバルをコピー

この場合、CLK_LOCAL_MEM_FENCEが必要になります

ここで問題に戻ります。私が見たところ、ワークグループ内の異なる項目がこの行を実行すると問題が発生する可能性があります。

w[i]-=h[i+j*n]*w[j];

同時にではありません。1つの作業項目がw[i]の値をすでに計算していて、他の作業項目がw[j]にアクセスするとします。次に、2番目の作業項目の「j」が最初の作業項目の「i」と同じである場合、他の作業項目は、最初の作業項目によってすでに更新されていた最初の反復値を使用します。

次に行うべきことは次のとおりです（グローバルメモリを使用したい場合）：

また、n <N（ワークグループのサイズ）と仮定します。そうしないと、複数のワークグループにまたがるため、同期できません。

for (j=0;j<n;j++)
{
    double wj;
    if (j<i)
        wj = w[j];
    barrier(CLK_GLOBAL_MEM_FENCE); // read_mem_fence(CLK_GLOBAL_MEM_FENCE) is enough
    if(j<i)
        w[i]-=h[i+j*n]*wj;
    barrier(CLK_GLOBAL_MEM_FENCE);  // write_mem_fence(CLK_GLOBAL_MEM_FENCE) is enough
}

お役に立てれば

score 0 · Accepted Answer

カーネル内で global_id と local_id の両方が必要な理由を詳しく教えてください。

ワークグループが 1 つしかない場合は、local_id で十分です。

また、なぜ g から w にデータをコピーするのですか?

単に w=h*g (h は行列、g はベクトル) 以上のことを達成しようとしていますか?

最後に、単にアプリケーションを複数回再起動するのではなく、単一のアプリケーションでカーネルを複数回起動するだけの場合、最も可能性の高い説明はどこかでメモリを破損しているようです。入力データを上書きしています。

カーネルに渡された入力データが同じ実行で一貫しているかどうかを確認できますか?

opencl - opencl 同期

2 に答える 2

Related

Reference