0

私のプログラムには、「aaaa」「bbbb」「cccc」などの 4 バイト文字列がたくさんあります... crc チェックに合格する特定の文字列を収集する必要があります。

文字列が crc チェックを通過できる可能性はほとんどないため、すべての結果を保持するために非常に大きなバッファーを使用したくありません。入力と同じように、結果を 1 つずつ連結することを好みます。たとえば、入力が「aaaabbbbcccc」で、「bbbb」が crc チェックに合格しない場合、出力文字列は「aaaacccc」で、output_count は 2 である必要があります。

コードは次のようになります。

__device__
bool is_crc_correct(char* str, int len) {
    return true; // for simplicity, just return 'true';
}

// arguments:
// input: a sequence of 4-bytes-string, eg: aaaabbbbccccdddd....
__global__
void func(char* input, int* output, int* output_count) {
    unsigned int index = blockDim.x*blockIdx.x + threadIdx.x;

    if(is_crc_correct(input + 4*index)) {
        // copy the string
        memcpy(output + (*output_count)*4,
               input + 4*index,
               4);
        // increase the counter
        (*output_count)++;
    }
}

明らかに、メモリ コピーはスレッド セーフではありません。atomicAdd 関数を ++ 操作に使用できることはわかっていますが、output と output_count の両方をスレッド セーフにする方法はありますか?

4

2 に答える 2

1

これを提案すると地獄に落ちるかもしれませんが、カーネル内でメモリを動的に割り当てるのはどうですか? 例については、この質問/回答を参照してください: CUDA は __device__ 関数でメモリを割り当てます

次に、共有メモリ配列を各カーネルに渡します。カーネルが実行された後、配列の各要素は、動的に割り当てられたメモリの一部または NULL を指します。したがって、スレッドブロックが実行された後、単一のスレッドで最終的なクリーンアップ カーネルを実行して、最終的な文字列を構築します。

于 2015-11-10T10:17:28.080 に答える