1

GPU GTX4651GBを搭載したMSVS2005でCUDASDK3.1を使用しています。私はそのようなカーネル関数を持っています：

__global__ void CRT_GPU_2(float *A, float *X, float *Y, float *Z, float *pIntensity, float *firstTime, float *pointsNumber)
{


  int holo_x = blockIdx.x*20 + threadIdx.x;
  int holo_y = blockIdx.y*20 + threadIdx.y;

  float k=2.0f*3.14f/0.000000054f;

  if (firstTime[0]==1.0f)
  {
   pIntensity[holo_x+holo_y*MAX_FINAL_X]=0.0f; 
  }

  for (int i=0; i<pointsNumber[0]; i++)
  {
   pIntensity[holo_x+holo_y*MAX_FINAL_X]=pIntensity[holo_x+holo_y*MAX_FINAL_X]+A[i]*cosf(k*sqrtf(pow(holo_x-X[i],2.0f)+pow(holo_y-Y[i],2.0f)+pow(Z[i],2.0f)));
  }

  __syncthreads(); 


}

これはカーネル関数を呼び出す関数です。

extern "C" void go2(float *pDATA, float *X, float *Y, float *Z, float *pIntensity, float *firstTime, float *pointsNumber)
{
 dim3 blockGridRows(MAX_FINAL_X/20,MAX_FINAL_Y/20);
 dim3 threadBlockRows(20, 20);

 CRT_GPU_2<<<blockGridRows, threadBlockRows>>>(pDATA, X, Y, Z, pIntensity,firstTime, pointsNumber); 
 CUT_CHECK_ERROR("multiplyNumbersGPU() execution failed\n");
 CUDA_SAFE_CALL( cudaThreadSynchronize() );
}

この関数のすべてのパラメーターをループでロードしています（たとえば、1回のループ反復でパラメーターごとに4096個の要素）。合計すると、すべてのループの反復後に、各パラメーターの32768要素に対してこのカーネルを作成したいと思います。

MAX_FINAL_Xは1920で、MAX_FINAL_Yは1080です。

alghoritmを開始すると、最初の反復が非常に速くなり、さらに1〜2回反復すると、CUDAタイムアウトエラーに関する情報が得られます。私はGPUgtx260でこのアルゴリズムを使用しましたが、覚えている限りではうまくいっていました...

私を助けてくれませんか..このアルゴリズムの新しいFermiアーチによると、私はいくつかの間違いをしているのでしょうか？

4

3 に答える 3

1

GPUはディスプレイに接続されていますか？もしそうなら、デフォルトでは、カーネルの実行は5秒後に中止されると思います。を使用して、カーネルの実行がタイムアウトするかどうかを確認できます。リファレンスページcudaGetDevicePropertiesを参照してください。

于 2010-07-12T07:31:58.140 に答える
1
  1. CUT_CHECK_ERROR後 に呼び出す方が良いでしょう cudaThreadSynchronize()。カーネルは非同期で実行され、カーネルがエラーを認識するまで待つ必要があるため... 2回目の反復で、最初のカーネルの使用からエラーが発生する可能性があります。
  2. 最も興味深い変数に有効な数値があることを確認してください pointsNumber[0]（長い内部ループが発生する可能性があります）。
  3. カーネル関数の速度を改善することもできます。
    • より良いブロックを使用してください。スレッド構成20x20では、メモリ使用量が非常に遅くなります（プログラミングガイドとベストプラクティスを参照）。ブロック16x16を使用してみてください。
    • 機能を使用しないでくださいpow(..., 2.0)。SQRマクロを使用する方が高速です（#define SQR(x) (x)*(x)）
    • 共有メモリを使用しないため、__syncthreads()必須ではありません。

PS：ポインターだけでなく、値パラメーターをCUDA関数に渡すこともできます。速度は同じになります。

PPS：コードの可読性を改善してください...ブロック構成を変更するには、6つの場所を編集する必要があります...カーネル内では、blockDim変数を使用でき、go2関数で定数を使用できます。また、使用することもできますbool firstTime-それよりもはるかに優れていfloatます。

于 2010-07-12T12:43:21.290 に答える
1

カーネルのサイクルでは、同じ配列に書き込み、そこから読み取ります。グローバルメモリ使用量の場合、異なるブロックからのワープが互いに待機するため、これは最悪です。

于 2012-04-14T08:56:58.587 に答える