CUDA线程通信

Question

我是CUDA编程的新手，我有点问题。我试图编写一个需要线程间通信的程序;我尝试了我找到的各种可能的方式，但仍然无法正常工作。你怎么看，我错过了什么？

下面的代码片段是我的整个程序。它在同一个块中启动2个线程。他们得到一个输入，一个输出数组，以及另一个全局变量来通信。值0表示该变量为空，因此可写。基本上，第一个从输入读取元素，将值传递给第二个元素，将其写入输出数组中。后来，它应该是一个管道，A和B.

#include <cuda.h> 
#include <cuda_runtime.h> 

#include <stdio.h> 

#define N 1 

__global__ void link(int *in, int *out, int *pipe){ 
    int id = threadIdx.y*blockDim.x + threadIdx.x; //compute index  

    if(id == 0){  //writer thread 

     for(int index = 0;index<N;){     
      if(pipe[0]==0){    
       atomicExch(pipe, in[index++]);    
      }   
     } 
    } 
    else if(id == 1){ // reader thread  

     for(int index=0;index<N;) { 
      if(pipe[0]!=0){ 
       out[index++] = atomicExch(pipe, 0); //read and make it empty  
      }   
     }   
    } 
} 

int main(){ 
    int input[] = {8,7}; 
    int *dev_input; 
    int *dev_output; 
    int *dev_pipe; 
    int *output = (int*) malloc (N*sizeof(int)); 

    cudaMalloc((void**) &dev_input, N*sizeof(int)); 
    cudaMalloc((void**) &dev_output, N*sizeof(int)); 
    cudaMalloc((void**) &dev_pipe, 1*sizeof(int)); 
    cudaMemset(dev_pipe, 0, 1); 
    cudaMemcpy(dev_input, &input[0], N*sizeof(int), cudaMemcpyHostToDevice); 

    link<<<1, 2>>>(dev_input, dev_output, dev_pipe); 

    cudaMemcpy(output, dev_output, N*sizeof(int), cudaMemcpyDeviceToHost); 

    printf("[%d", output[0]); 
    for(int i = 1;i<N;i++) 
     printf(", %d", output[i]); 
    printf("]\n"); 
    int d = 0; 
    scanf("\n", &d); 

} 

如果读者看到之间更多的线程，该管道是0（空），提出的第一个元素就可以了，但笔者不能看到任何更改，程序进入僵局。我试图添加__threadfence和__syncthreads，但它没有帮助。我也尝试了volatile共享内存，但它也没有工作。请帮助我，如果可以的话，因为我不知道，它有什么问题。

Answer 1

要小心，CUDA线程与POSIX线程有很大不同。它们按照单指令多线程模式工作（SIMT，参见this interesting discussion）：在每个时钟周期，每个线程（在一个'wrap'中）运行相同的（低级）指令。

在您的代码中，编写程序线程将在读取程​​序线程执行时运行，然后，第二个线程将在第一个线程执行时运行，但它们将永远不会同时运行，因此您不会从大量GPU的并行结构。

总之，要回答你的问题，你for循环

for(int index=0;index<N;) 

不增加指标;所以他们是无限循环。与

for(int index=0;index<N;index++) 

更换和共享内存，这是常见的相同的块内的所有线程，会比对线程间通信全局内存快很多。