用于简单计算的 CUDA 加速

Question

我在 cuda_computation.cu 中有以下代码

#include <iostream>
#include <stdio.h>
#include <cuda.h>
#include <assert.h>

void checkCUDAError(const char *msg);

__global__ void euclid_kernel(float *x, float* y, float* f)
{
  int idx = blockIdx.x*blockDim.x + threadIdx.x;
  int i = blockIdx.x;
  int j = threadIdx.x;
  f[idx] = sqrt((x[i]-x[j])*(x[i]-x[j]) + (y[i]-y[j])*(y[i]-y[j]));
}
int main()
{
  float *xh;
  float *yh;
  float *fh;
  float *xd;
  float *yd;
  float *fd;

  size_t n = 256;
  size_t numBlocks = n;
  size_t numThreadsPerBlock = n;

  size_t memSize = numBlocks * numThreadsPerBlock * sizeof(float);
  xh = (float *) malloc(n * sizeof(float));
  yh = (float *) malloc(n * sizeof(float));
  fh = (float *) malloc(memSize);

  for(int ii(0); ii!=n; ++ii)
    {
      xh[ii] = ii;
      yh[ii] = ii;
    }

  cudaMalloc( (void **) &xd, n * sizeof(float) );
  cudaMalloc( (void **) &yd, n * sizeof(float) );
  cudaMalloc( (void **) &fd, memSize );
  for(int run(0); run!=10000; ++run)
    {
      //change value to avoid optimizations
      xh[0] = ((float)run)/10000.0;
      cudaMemcpy( xd, xh, n * sizeof(float), cudaMemcpyHostToDevice );
      checkCUDAError("cudaMemcpy");
      cudaMemcpy( yd, yh, n * sizeof(float), cudaMemcpyHostToDevice );
      checkCUDAError("cudaMemcpy");
      dim3 dimGrid(numBlocks);
      dim3 dimBlock(numThreadsPerBlock);
      euclid_kernel<<< dimGrid, dimBlock >>>( xd, yd, fd );
      cudaThreadSynchronize();
      checkCUDAError("kernel execution");
      cudaMemcpy( fh, fd, memSize, cudaMemcpyDeviceToHost );
      checkCUDAError("cudaMemcpy");
    }
  cudaFree(xd);
  cudaFree(yd);
  cudaFree(fd);
  free(xh);
  free(yh);
  free(fh);
  return 0;
}

void checkCUDAError(const char *msg)
{
  cudaError_t err = cudaGetLastError();
  if( cudaSuccess != err) 
    {
      fprintf(stderr, "Cuda error: %s: %s.\n", msg, cudaGetErrorString( err) );
      exit(-1);
    }                         
}

在 FX QUADRO 380 上运行大约需要 6"，而仅使用一个 i7-870 内核的相应串行版本只需大约 3"。我想念什么吗？代码是否在某些方面进行了优化？或者对于简单的计算(例如这种全对欧几里德距离)，移动内存所需的开销是否超过了计算增益，这只是预期的行为吗？

Answer 1

我认为您在移动数据时被杀死了。尤其是因为您使用单个值调用 CUDA 内核，所以将大量值作为一维数组上传并对其进行操作可能会更快。

而且 sqrt 没有在 Cuda 上的 HW 中完成(至少在我的 GPU 上没有)，而 CPU 已经为此优化了 FPU HW，并且可能比 GPU 快 10 倍，对于像这样的小工作可能会保留所有在 timign 运行之间的缓存中的结果。