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

Posted 2023-04-15

【中文标题】CUDA 中的分歧 - 从内核中的线程退出

【英文标题】：Divergence in CUDA - exit from a thread in kernel

【发布时间】：2013-01-29 22:35:24

【问题描述】：

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

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

或

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

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

【问题讨论】：

两者都会给你同样的表现。

【参考方案1】：

虽然两者在性能上是一样的，但是你应该考虑到第一个是不推荐的。

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

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

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

但是在 __syncthreads() 的情况下，结果有点悲惨。 CUDA 架构保证在块中的每个线程都执行完 __syncthreads() 之前，没有线程会前进到超出 __syncthreads() 的指令

所以，它主要与内核中的线程同步有关。你可以在这里找到关于这个主题的一个很好的问题/答案：Can I use __syncthreads() after having dropped threads?

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

【讨论】：

意外行为是什么意思，有什么问题？我在教程中看到过它，我也使用过它，到目前为止没有出现任何问题。

请为您的声明提供参考。

@GeorgeAprilis 问题主要与良好实践和未来块内同步有关。

正如@RoBik 建议的那样，我提供了一些参考资料。