2

CUDA プログラミング ガイドのセクション C.4.3.1.2 の下。「ネスティングと同期の深さ」には、次のように記載されています。

「親カーネルが cudaDeviceSynchronize() を呼び出さない場合に、親の状態用にスペースを予約する必要がないことをシステムが検出した場合、最適化が許可されます。この場合、明示的な親/子同期が発生しないため、プログラムは保守的な最大値よりもはるかに少なくなります。そのようなプログラムは、バッキング ストアの過剰な割り当てを避けるために、より浅い最大同期深度を指定できます。"

これは、コンパイラが動的並列処理で末尾再帰をサポートしていることを意味しますか? たとえば、自分自身を再帰的に呼び出すカーネルがあるとします。

__global__ void kernel(int layer){
  if(layer>65535){
    return;
  }
  printf("layer=%d\n",layer);
  kernel<<<1,1>>>(layer+1);
}

ホストで起動:

   kernel<<<1,1>>>(0);

末尾再帰をコンパイラで最適化できる場合、「親/子の同期は発生しない」ため、最大再帰レベル 24 に制限されますか? 制限されていない場合、コンパイラはどのように最適化を有効にできますか?

ありがとう!

4

0 に答える 0