[解決済み] cuda atomicAdd のサンプルで正しい出力が得られない

質問

以下のコードは、100要素のfloatの配列を10回1ずつインクリメントすることを目的に書かれたものです。出力では、各要素に10.0fの値を持つ100要素の配列を期待していました。その代わり、ランダムな値が出力されます。ここで私の誤りを指摘してもらえますか？

__global__  void testAdd(float *a)
{
    float temp;
    for (int i = 0; i < 100 ; i++)
    {
        a[i] = atomicAdd(&a[i], 1.0f);
    }
}
void cuTestAtomicAdd(float *a)
{
    testAdd<<<1, 10>>>(a);
}

アトム演算の仕組みを理解し、他の分野にも応用できるようになることが目標です。

どのように解決するのか？

それは atomicAdd 演算を行うことができます。

こんな感じでやればいいんです。

atomicAdd(&a[i], 1.0f);

で、当該変数( a[i] ) が更新されます。

は 戻り値 アトム関数から得られるのは、一般に 古い の値は、変数にあったものです。 以前 アトミックな更新を行います。

ということで、こうすることで

a[i] = atomicAdd(&a[i], 1.0f);

は変数を更新します。 a[i] を指定し、（解剖学的でない方法で）その後に 古い の値を変数 a[i] . これは、ほぼ間違いなく、あなたが望むものではありません。

を読んでみてください。 ドキュメント :

この関数は old を返します。

次の完全なコードは、正しい使い方を示しています。

#include <iostream>

__global__  void testAdd(float *a)
{
    for (int i = 0; i < 100 ; i++)
    {
        atomicAdd(&a[i], 1.0f);
    }
}
void cuTestAtomicAdd(float *a)
{
    testAdd<<<1, 10>>>(a);
}

int main(){

  float *d_data, *h_data;
  h_data=(float *) malloc(100*sizeof(float));
  cudaMalloc((void **)&d_data, 100*sizeof(float));
  cudaMemset(d_data, 0, 100*sizeof(float));
  cuTestAtomicAdd(d_data);
  cudaMemcpy(h_data, d_data, 100*sizeof(float), cudaMemcpyDeviceToHost);
  for (int i = 0; i < 100; i++)
    if (h_data[i] != 10.0f) {printf("mismatch at %d, was %f, should be %f\n", i, h_data[i], 10.0f); return 1;}
  printf("Success\n");
  return 0;
}