performance - CUDA での発散 - カーネルでのスレッドからの終了

Question

スレッドインデックスが大きすぎるスレッドを終了するにはどうすればよいでしょうか。2 つの可能性があります。

int i = threadIdx.x;
if(i >= count)
    return;
// do logic

また

int i = threadIdx.x;
if(i < count) {
    // do logic
}

どちらも正しいことはわかっていますが、パフォーマンスに影響を与えるのはどちらですか?

score 4 · Accepted Answer

どちらもパフォーマンスの点では同じですが、最初のものは推奨されないことを考慮する必要があります。

カーネル内でスレッドを返すと、残りのコードで予期しない動作が発生する可能性があります。

予期しない動作とは、ワープにグループ化されたスレッドの最小単位に関連する問題を意味します。たとえば、カーネルにブロックがある場合if / else、この状況はスレッド分岐と呼ばれ、通常の場合、スレッドがアイドル状態のままになり、他のスレッドがいくつかの命令を実行します。

CUDA by Example Book、第 5 章、スレッドの協力:

しかし、__syncthreads() の場合、結果はいくぶん悲劇的です。CUDA アーキテクチャは、ブロック内のすべてのスレッドが __syncthreads() を実行するまで、スレッドが __syncthreads() を超えて命令に進まないことを保証します。

そのため、主にカーネル内のスレッド同期に関連しています。ここで、このトピックに関する非常に良い質問/回答を見つけることができます:スレッドを削除した後に __syncthreads() を使用できますか?

最後のメモとして、私もその悪い習慣を使用しましたが、問題は発生しませんでしたが、将来問題が発生する可能性があるという保証はありません。あまりお勧めできない内容です

1 に答える 1