CUDA 中的原子操作？要包含哪个头文件？

Question

在 CUDA 中使用原子操作，是否需要包含一些 CUDA 头文件？ CUDA 编程指南似乎对此守口如瓶。

下面给出的代码 glmax.cu 给了我以下编译错误。

gaurish108 MyPractice: nvcc glmax.cu -o glmax
glmax.cu(11): error: identifier "atomicMax" is undefined

1 error detected in the compilation of "/tmp/tmpxft_000010fa_00000000-4_glmax.cpp1.ii".

这里是代码。它基本上是使用原子操作atomicMax 在 GPU 上计算数组的最大值。因为我是 CUDA 的新手，所以我确信这是一个非常幼稚的代码，但我写这个是为了帮助自己理解原子操作。

#include<stdio.h>
#include<stdlib.h>
#include<math.h>

__global__ void global_max(int* values, int* gl_max)
{

  int i=threadIdx.x + blockDim.x * blockIdx.x;
  int val=values[i];

  atomicMax(gl_max,val);

}


int main(void)
{
  int array_size=5;
  int num_bytes=array_size*sizeof(int);
  int *device_array=0;
  int *host_array=0;

  int *device_max=0;
  int *host_max=0;

  //Allocate memory on the host
  host_array=(int*)malloc(num_bytes);

  //Allocate memory on the device
  cudaMalloc((void**)&device_array,num_bytes);
  cudaMalloc((void**)&device_max,sizeof(int));


  //If either memory allocation failed, report an error message
  if(host_array == 0 || device_array == 0)
  {
    printf("couldn't allocate memory\n");
    return 1;
  }

  //Assign a random integer in the  interval [0,25] to host_array members
  for(int i=0;i<array_size;++i)
    {
      *(host_array+i)=rand()%26;
    }

  //Print the host array members
  printf("Host Array\n");
  for(int i=0;i<array_size;++i)
    {
      printf("%d  ",*(host_array+i));
    }
  printf("\n");

  //Copy array from host to device.
  cudaMemcpy(device_array,host_array,num_bytes,cudaMemcpyHostToDevice);

  //Configure and launch the kernel which calculates the maximum element in the device array.
  int grid_size=1;//Only 1 block of threads is used
  int block_size=5;//One block contains only 5 threads

  //Device array passed to the kernel as data. 
  global_max<<<grid_size,block_size>>>(device_array,device_max);

  //Transfer the maximum value so calculated into the CPU and print it
  cudaMemcpy(host_max,device_max,sizeof(int),cudaMemcpyDeviceToHost);
  printf("\nMaximum value is %d\n",*host_max);


  // deallocate memory
  free(host_array);
  cudaFree(device_array);
  cudaFree(device_max);
  return 0;
}

Answer 1

我不认为#include 是必要的。原子操作在“计算能力”1.0 (sm_10) 设备上不可用，这是您要求 nvcc 编译的（默认情况下）。

要在代码中使用atomicMax，请在命令行中至少指定-arch=sm_11：

$nvcc -arch=sm_11 glmax.cu -o glmax

为了将来参考，您可以查阅 CUDA C 编程指南的附录 F，了解在特定计算能力的平台上可以使用哪些原子操作。

当然，您需要sm_11 兼容的 GPU 才能执行代码。我的印象是这些现在很常见。