1

cl_khr_int64_base_atomics をサポートしない ATI Firepro V4800 グラフィック カードを使用しています。RadixSort アルゴを長い整数に適応させようとしています。アルゴは、64 ビット版の atom_inc をカーネルで使用できませんが、atomic_inc を使用します。だから、私の質問は、使用できるatomic_incと同じ機能を実行するコードがありますか? カーネル コードの一部を以下に示します。

__kernel void histogram(__global uint* unsortedData,
           __global uint* buckets,
           uint shiftCount,
            __local uint* sharedArray)
{
    size_t localId = get_local_id(0);
    size_t globalId = get_global_id(0);
    size_t groupId = get_group_id(0);
    size_t groupSize = get_local_size(0);

    uint numGroups = get_global_size(0) / get_local_size(0);


    // Initialize shared array to zero //

    sharedArray[localId] = 0;

    barrier(CLK_LOCAL_MEM_FENCE);

   // Calculate thread-histograms //
    uint value = unsortedData[globalId];
  value =  value >> shiftCount & 0xFFU;    
    atomic_inc(sharedArray+value);


    barrier(CLK_LOCAL_MEM_FENCE);
    // Copy calculated histogram bin to global memory //

    uint bucketPos = groupId  * groupSize + localId ;
    //uint bucketPos = localId * numGroups + groupId ;
    buckets[bucketPos] = sharedArray[localId];
}

助言がありますか?ありがとうございました。

編集: 同じことの別の方法は、このブログサイトに記載されています: http://suhorukov.blogspot.in/2011/12/opencl-11-atomic-operations-on-floating.html . これにより、Atomic Inc. の非常に汎用的な実装が得られます。

4

1 に答える 1