__syncthreads() Deadlock

本文关键字：Deadlock syncthreads 更新时间：2023-10-16

如果只有一些线程执行__syncthreads（），它会导致死锁吗？

我有一个像这样的内核：

__global__ void Kernel(int N,int *a)
{
    if(threadIdx.x<N)
    {
      for(int i=0;i<N;i++)
       {
        a[threadIdx.x]= //Some calculation using a and i
        __syncthreads()
       }
    }
}

如果块中的线程数大于N，则某些线程将不会执行代码。这会造成僵局吗
如果是，那么我如何修改代码？

您应该不要在发散代码中使用__syncthreads()。它在这种情况下的行为是不明确的。

__syncthreads()可能出现在条件分支中，前提是您确信该分支将以相同的方式由块中的所有线程统一求值（块中的全部线程或没有线程使用该分支）。

从技术上讲，内核不能死锁；他们只能暂停。但是，是的，你所描述的是真实的，是可能发生的。这个问题以前已经讨论过了，例如：CUDA/OpenCL 中的真实死锁示例