4

私のプログラムでは、ボクセル グリッドを介して多数の粒子を追跡しています。ボクセルに対するパーティクルの比率は任意です。ある時点で、どのパーティクルがどのボクセルにあり、いくつあるかを知る必要があります。具体的には、ボクセルは、どの粒子が含まれているかを正確に認識している必要があります。CUDA のようなものは使用できないためstd::vector、次のアルゴリズムを (高レベルで) 使用しています。

  • ボクセル数のサイズの int の配列を割り当てます
  • すべてのパーティクルのスレッドを起動し、各パーティクルが存在するボクセルを特定し、「バケット」配列の適切なカウンターを増やします
  • 粒子の数のサイズのポインターの配列を割り当てます
  • この新しい配列への各ボクセルのオフセットを計算します (その前のボクセル内の粒子の数を合計します)
  • パーティクルを配列内に順序どおりに配置します (このデータを使用して、後で操作を高速化します。速度の向上は、メモリ使用量の増加に見合うだけの価値があります)。

ただし、これは2番目のステップで崩壊します。私は長い間 CUDA でプログラミングを行っていませんが、スレッド間でグローバル メモリ内の同じ場所に同時に書き込むと、未定義の結果が生じることがわかりました。これは、ほとんどの場合 で 1 を取得し、ときどき 2 を取得するという事実に反映されていbucketsます。このステップで使用しているコードのスケッチを次に示します。

__global__ void GPU_AssignParticles(Particle* particles, Voxel* voxels, int* buckets) {
    int tid = threadIdx.x + blockIdx.x*blockDim.x;

    if(tid < num_particles) { // <-- you can assume I actually passed this to the function :)
        // Some math to determine the index of the voxel which this particle
        // resides in.
        buckets[index] += 1;
    }
}

私の質問は、CUDA でこれらのカウントを生成する適切な方法は何ですか?

また、ボクセル内のパーティクルへの参照を保存する方法はありますか? 私が目にする問題は、ボクセル内の粒子の数が絶えず変化することです。そのため、新しい配列はほぼすべてのフレームで割り当てを解除して再割り当てする必要があります。

4

1 に答える 1

1

バケット数を計算するためのより効率的なソリューションがあるかもしれませんが、最初の実用的なソリューションは、現在のアプローチを使用することですが、アトミックインクリメントを使用します。このように、一度に1つのスレッドのみがバケット数をアトミックに増分します（グリッド全体で同期されます）。

if(tid < num_particles) {
    // ...
    atomicAdd(&buckets[index], 1);
}
于 2012-04-13T08:32:03.527 に答える