ヒュージグラムによる計算の計算

Question

コードは機能しません。しかし、次のコードでatomicAddとコメントすると、コードが機能します。

その理由は何ですか？
フロート配列のヒストグラムコードはどこで入手できますか？

__global__ void calculateHistogram(float *devD, int* retHis) 
    { 
     int globalLi = getCurrentThread(); //get the thread ID 
     if(globalLi>=0 && globalLi<Rd*Cd*Dd) 
     { 
      int r=0,c=0,d=0; 
      GetInd2Sub(globalLi, Rd, Cd, r, c, d); //some calculations to get r,c,d 
      if(r>=stYd && r<edYd && c>=stXd && c<edXd && d>=stZd && d<edZd) 
      { 
      //calculate the histogram 
      int indexInHis = GetBinNo(devD[globalLi]); //get the bin number in the histogram 
      atomicAdd(&retHis[indexInHis],1); //when I comment this line the code works 
      } 
     } 
    } 

Answer 1

ジェイソン・サンダースとエドワードKandrotによって例によってCUDAの第9章を見てみましょう。アトミックをカバーし、8ビット整数のヒストグラムを計算する簡単な例を示します。最初のバージョンでは、各値に対してアトミック・アドレスが使用されていますが、動作しますが非常に遅いです。洗練されたこの例では、共有メモリ内の各ブロックのヒストグラムを計算し、すべてのヒストグラムをグローバルメモリにマージして最終結果を得ます。あなたのコードは最初のバージョンと似ています。いったん動作させると、より洗練されたバージョンのようになります。

あなたは両方のバージョンを確認するために本から例をダウンロードすることができます。CUDA by Example downloads

あなたは、完全なコードやエラーメッセージを与えることは表示されませんので、私はあなたのコードで間違っ起こっている正確に何と言うことはできません。ここではいくつかの考えです：

• あなたは（デフォルト1.0アーキテクチャの目標よりもすなわち大きい）アトミックをサポートアーキテクチャでコンパイルする必要があり
• は、インデックスとインデックスの制限が多少複雑に見える、私はそれらの
をダブルチェックします
• あなたのビンの計算はretHisの有効範囲外のビン番号を与えるかもしれない、私は少なくとも