CUDAカーネルが__device__関数を呼び出す場合、__ device__関数に出入りするブロック内のすべてのスレッド間に暗黙の同期がありますか?
そうでない場合は、ブロック内の他のスレッドが(明示的な同期がない場合に)入る前に、ブロック内の一部のスレッドが__device__関数を終了した可能性があることを意味しますか?
関連情報/参照へのポインタをいただければ幸いです。
CUDAカーネルが__device__関数を呼び出す場合、__ device__関数に出入りするブロック内のすべてのスレッド間に暗黙の同期がありますか?
そうでない場合は、ブロック内の他のスレッドが(明示的な同期がない場合に)入る前に、ブロック内の一部のスレッドが__device__関数を終了した可能性があることを意味しますか?
関連情報/参照へのポインタをいただければ幸いです。
はい、タロンミーが指摘したように、ワープには暗黙の同期btwスレッドしかありません。カーネルが起動されると、ハードウェアはワープ(おそらく最初のワープ)をピークし、その最初の命令を実行してから、別のワープに切り替えます。ワープは終了時に互いに待機する必要がないため、別のワープがカーネルの最初の命令を実行する前に、一部のワープがリタイアする可能性があります。
__syncthreads()
カーネルコードで関数を使用する必要があります。関数を呼び出した後、バリアを配置してスレッドを同期する行を__device__
追加します。__syncthreads();