1

サンプルから取得したコード。それを使ってプロジェクトを作成しましたが、うまくいきましたが、一部がわかりません。

例として、32x32のマトリックスがあり、36の作業項目があるため、get_global_id（0）は0-> 35と推定され、サイズ= MATRIX_DIM / 4=8になります。

__kernel void transpose(__global float4 *g_mat,
   __local float4 *l_mat, uint size) {

   __global float4 *src, *dst;

   /* Determine row and column location */
   int col = get_global_id(0);
   int row = 0;
   while(col >= size) {
      col -= size--;
      row++;
   }
   col += row;
   size += row;

   /* Read source block into local memory */
   src = g_mat + row * size * 4 + col;
   l_mat += get_local_id(0)*8;

clEnqueueNDRangeKernel呼び出しでは、argがlocal_work_size設定さNULLれ、手動の手段によれば、コンパイラまたは何かがそれを理解できるようにします。

local_work_size can also be a NULL value in which case the OpenCL implementation will determine how to be break the global work-items into appropriate work-group instances.

しかし、私は8を掛けることを理解していません。これは、私が想定しているワークグループのローカルメモリにアドレスオフセットを与えます。誰かがこれを説明できますか？

   l_mat[0] = src[0];
   l_mat[1] = src[size];
   l_mat[2] = src[2*size];
   l_mat[3] = src[3*size];

   /* Process block on diagonal */
   if(row == col) {
      src[0] =
         (float4)(l_mat[0].x, l_mat[1].x, l_mat[2].x, l_mat[3].x);
      src[size] =
         (float4)(l_mat[0].y, l_mat[1].y, l_mat[2].y, l_mat[3].y);
      src[2*size] =
         (float4)(l_mat[0].z, l_mat[1].z, l_mat[2].z, l_mat[3].z);
      src[3*size] =
         (float4)(l_mat[0].w, l_mat[1].w, l_mat[2].w, l_mat[3].w);
   }
   /* Process block off diagonal */
   else {
      /* Read destination block into local memory */
      dst = g_mat + col * size * 4 + row;
      l_mat[4] = dst[0];
      l_mat[5] = dst[size];
      l_mat[6] = dst[2*size];
      l_mat[7] = dst[3*size];

      /* Set elements of destination block */
      dst[0] =
         (float4)(l_mat[0].x, l_mat[1].x, l_mat[2].x, l_mat[3].x);
      dst[size] =
         (float4)(l_mat[0].y, l_mat[1].y, l_mat[2].y, l_mat[3].y);
      dst[2*size] =
         (float4)(l_mat[0].z, l_mat[1].z, l_mat[2].z, l_mat[3].z);
      dst[3*size] =
         (float4)(l_mat[0].w, l_mat[1].w, l_mat[2].w, l_mat[3].w);

      /* Set elements of source block */
      src[0] =
         (float4)(l_mat[4].x, l_mat[5].x, l_mat[6].x, l_mat[7].x);
      src[size] =
         (float4)(l_mat[4].y, l_mat[5].y, l_mat[6].y, l_mat[7].y);
      src[2*size] =
         (float4)(l_mat[4].z, l_mat[5].z, l_mat[6].z, l_mat[7].z);
      src[3*size] =
         (float4)(l_mat[4].w, l_mat[5].w, l_mat[6].w, l_mat[7].w);
   }
}
4

2 に答える 2

1

l_matは、すべての作業項目を反転する2つの行列コンポーネントを格納するための一時バッファーとして使用されます。

したがって、各作業項目について、 2 * 4のfloat4を格納する必要があります。したがって、offset = get_local_id（0）* 2 * 4 = get_local_id（0）*8です。

于 2012-11-07T18:20:32.940 に答える
1

l_matワークグループ内のスレッドのローカルストアで使用されています。具体的には、ローカルメモリへのアクセスがグローバルメモリよりも桁違いに速いために使用されています。

各スレッドには8float4秒必要です。次のポインタ演算を行う

l_mat += get_local_id(0)*8;

l_mat他のスレッドのデータと重複しないように、各スレッド のポインターを移動します。

local_sizeが指定されておらず、のサイズが各スレッドの値を格納するのに十分であることを確認できないため、これによりエラーが発生する可能性があります。l_mat

于 2012-11-07T18:25:32.310 に答える