0

親カーネルと子カーネルがある次のコードを見てみましょう。threadIdx.x上記の親カーネルから、並列スループットを最大化するために、異なるストリームで子カーネルを開始したいと考えています。cudaDeviceSynchronize()次に、親カーネルがメモリに加えられた変更を確認する必要があるため、それらの子を待ちglobalます。

ここnで、ストリームを使用して親カーネルを開始したいとします。また、並行して開始したい親カーネルの各セット間でn、次を使用して結果を待つ必要があります。cudaDeviceSynchronize()

これはどのように動作しますか?

このNvidiaによる動的並列処理の公式紹介からparent kernel[0]ストリームがその中で開始されるのを待つだけだと思います。これは正しいです?そうでない場合、どうなりますか?

注:一度に実行できるストリームの数は非常に多い(私の場合は32)ことは承知していますが、これは占有率を最大化するためのものです

編集:小さなコードサンプル

__global__ void child_kernel (void) {}
__global__ void parent_kernel (void) 
{
    if (blockIdx.x == 0)
    {
        cudaStream_t s;
        cudaStreamCreateWithFlags(&s, cudaStreamNonBlocking);

        child_kernel <<<1,10,0,s>>> ();
        cudaStreamDestroy(s);
    }
    cudaDeviceSynchronize();
}
for (int i=0; i<10; i++)
{
    cudaStream_t s;
    cudaStreamCreateWithFlags(&s, cudaStreamNonBlocking);

    parent_kernel <<<10,10,0,s>>> ();
    cudaStreamDestroy(s);
}
cudaDeviceSynchronize();
4

1 に答える 1