performance - CUDA 中的分歧 - 从内核中的线程退出

Question

我想知道如何退出线程索引太大的线程。我看到两种可能性:

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

或者

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

我知道，两者都是正确的，但哪一个对性能影响更大？

Answer 1

尽管两者在性能方面相同，但您应该考虑到不推荐使用第一个。

返回内核中的线程可能会导致其余代码出现意外行为。

我所说的意外行为是指与分组在一个扭曲中的线程的最小单位相关的任何问题。例如，如果您有 if / else在您的内核中阻塞，这种情况被称为线程分歧，在正常情况下，它会导致线程保持空闲，而其他线程执行一些指令。

CUDA by Example Book，第 5 章，线程协作:

But in the case of __syncthreads(), the result is somewhat tragic. The CUDA Architecture guarantees that no thread will advance to an instruction beyond the __syncthreads() until every thread in the block has executed the __syncthreads()

因此，它主要与内核中的线程同步有关。您可以在此处找到有关此主题的非常好的问题/答案: Can I use __syncthreads() after having dropped threads?

正如我最后指出的，我也使用了这种不好的做法，没有出现问题，但不能保证将来可能会出现问题。这是我不推荐的东西