如何从CUDA核函数中返回单个变量？

Question

我有一个CUDA搜索函数，它计算一个变量。如何将其返回？

__global__ 
void G_SearchByNameID(node* Node, long nodeCount, long start,char* dest, long answer){
    answer = 2;
}

cudaMemcpy(h_answer, d_answer, sizeof(long), cudaMemcpyDeviceToHost);
cudaFree(d_answer);

对于这两行代码，我都得到了以下错误： 错误：类型为“long”的参数与类型为“const void *”的参数不兼容

- Pouya BCD

Answer 1

我一直在使用__device__变量来达到这个目的，这样你就不必费心去处理cudaMalloc和cudaFree，也不必将指针作为内核参数传递，这样可以节省内核中的寄存器。

__device__ long d_answer;

__global__ void G_SearchByNameID() {
  d_answer = 2;
}

int main() {
  SearchByNameID<<<1,1>>>();
  typeof(d_answer) answer;
  cudaMemcpyFromSymbol(&answer, "d_answer", sizeof(answer), 0, cudaMemcpyDeviceToHost);
  printf("answer: %d\n", answer);
  return 0;
}

- wich

---

@Erogol 内核和主机代码都从 __device__ 声明中了解类型。 - wich

由于某种原因，在我的情况下（工具包6.5），只有当我将“d_answer”替换为d_answer，即删除引号时，它才起作用。除此之外，这个方法很有效。 - icurays1

对我来说，该值始终返回为零。 - orodbhen

一个 __device__ 变量可以是线程本地的吗？像那样的全局变量对于多线程应用程序不起作用。 - Serge Rogatch

@Serge，老实说我不知道，我从来没有尝试过多线程应用程序。 - wich

这种方法不是“可重入”的，不能支持从多个流中启动相同内核。 - einpoklum

Answer 2

要获取单个结果，您需要使用Memcpy，即：

#include <assert.h>

__global__ void g_singleAnswer(long* answer){ *answer = 2; }

int main(){

  long h_answer;
  long* d_answer;
  cudaMalloc(&d_answer, sizeof(long));
  g_singleAnswer<<<1,1>>>(d_answer);
  cudaMemcpy(&h_answer, d_answer, sizeof(long), cudaMemcpyDeviceToHost); 
  cudaFree(d_answer);
  assert(h_answer == 2);
  return 0;
}

我猜测错误是由于您传递了一个长整型的值，而不是长整型值的指针导致的。

- fabrizioM

---

你不一定需要使用memcpy()函数 - 还有其他选择，比如@wich的回答。除非你认为复制是最好的/唯一现实的选择，否则请勿使用-1。 - einpoklum

请注意，另一个选项确实会执行memcpy操作（它在名称cudaMemcpyFromSymbol中）。这个答案使用了动态分配，相对于使用全局变量的另一个选项更适用于多线程应用程序。 - rgov