使用CUDA运行时API检查错误的规范方法是什么？

Question

查看有关CUDA问题的答案和评论,以及CUDA标记维基,我发现通常建议每个API调用的返回状态都应该检查错误.API文档包括像功能cudaGetLastError,cudaPeekAtLastError以及cudaGetErrorString,但什么是把这些结合在一起,以可靠地捕捉和无需大量额外的代码报告错误的最好方法？

Answer 1

检查运行时API代码中的错误的最佳方法可能是定义一个断言样式处理函数和包装器宏,如下所示:

#define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
{
   if (code != cudaSuccess) 
   {
      fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
      if (abort) exit(code);
   }
}

然后你可以用与每个API调用gpuErrchk宏,将处理的API的返回状态称之为包装,例如:

gpuErrchk( cudaMalloc(&a_d, size*sizeof(int)) );

如果调用中出现错误,将发出描述错误的文本消息以及发生错误的代码中的文件和行,stderr应用程序将退出.你可以设想修改gpuAssert引发异常,而不是exit()在需要时调用更复杂的应用程序.

第二个相关的问题是如何为您在内核启动的错误,它不能直接包装在如标准运行时API调用宏调用.对于内核,这样的事情:

kernel<<<1,1>>>(a);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaDeviceSynchronize() );

将首先检查无效的启动参数,然后强制主机等待内核停止并检查执行错误.如果您有后续阻塞API调用,则可以消除同步:

kernel<<<1,1>>>(a_d);
gpuErrchk( cudaPeekAtLastError() );
gpuErrchk( cudaMemcpy(a_h, a_d, size * sizeof(int), cudaMemcpyDeviceToHost) );

在这种情况下,cudaMemcpy调用可以返回在内核执行期间发生的错误或来自内存复制本身的错误.这对初学者来说可能会造成混淆,我建议在调试期间内核启动后使用显式同步,以便更容易理解可能出现问题的位置.

Answer 2

上面的talonmies回答是以一种方式中止应用程序的好assert方法.

有时,我们可能希望报告并从C++上下文中的错误条件中恢复,作为更大应用程序的一部分.

通过抛出std::runtime_error使用thrust::system_error以下派生的C++异常,这是一种相当简洁的方法:

#include <thrust/system_error.h>
#include <thrust/system/cuda/error.h>
#include <sstream>

void throw_on_cuda_error(cudaError_t code, const char *file, int line)
{
  if(code != cudaSuccess)
  {
    std::stringstream ss;
    ss << file << "(" << line << ")";
    std::string file_and_line;
    ss >> file_and_line;
    throw thrust::system_error(code, thrust::cuda_category(), file_and_line);
  }
}

这将把文件名,行号和英语语言描述cudaError_t合并到抛出的异常.what()成员中:

#include <iostream>

int main()
{
  try
  {
    // do something crazy
    throw_on_cuda_error(cudaSetDevice(-1), __FILE__, __LINE__);
  }
  catch(thrust::system_error &e)
  {
    std::cerr << "CUDA error after cudaSetDevice: " << e.what() << std::endl;

    // oops, recover
    cudaSetDevice(0);
  }

  return 0;
}

输出:

$ nvcc exception.cu -run
CUDA error after cudaSetDevice: exception.cu(23): invalid device ordinal

如果需要,客户端some_function可以将CUDA错误与其他类型的错误区分开来:

try
{
  // call some_function which may throw something
  some_function();
}
catch(thrust::system_error &e)
{
  std::cerr << "CUDA error during some_function: " << e.what() << std::endl;
}
catch(std::bad_alloc &e)
{
  std::cerr << "Bad memory allocation during some_function: " << e.what() << std::endl;
}
catch(std::runtime_error &e)
{
  std::cerr << "Runtime error during some_function: " << e.what() << std::endl;
}
catch(...)
{
  std::cerr << "Some other kind of error during some_function" << std::endl;

  // no idea what to do, so just rethrow the exception
  throw;
}

因为thrust::system_error是a std::runtime_error,如果我们不需要前一个例子的精度,我们也可以用相同的广义错误类型来处理它:

try
{
  // call some_function which may throw something
  some_function();
}
catch(std::runtime_error &e)
{
  std::cerr << "Runtime error during some_function: " << e.what() << std::endl;
}

Answer 3

C++ - 规范方式:不检查错误...使用抛出异常的C++绑定.

我曾经对这个问题感到厌烦; 而且我曾经有一个宏观兼容包装功能的解决方案,就像在Talonmies和Jared的答案中一样,但老实说呢？它使得使用CUDA Runtime API变得更加丑陋和类似C语言.

所以我以一种不同的,更基本的方式接近了这一点.有关结果的示例,这里是CUDA vectorAdd示例的一部分- 对每个运行时API调用进行完整的错误检查:

// (... prepare host-side buffers here ...)

auto current_device = cuda::device::current::get();
auto d_A = cuda::memory::device::make_unique<float[]>(current_device, numElements);
auto d_B = cuda::memory::device::make_unique<float[]>(current_device, numElements);
auto d_C = cuda::memory::device::make_unique<float[]>(current_device, numElements);

cuda::memory::copy(d_A.get(), h_A.get(), size);
cuda::memory::copy(d_B.get(), h_B.get(), size);

// (... prepare a launch configuration here... )
cuda::launch( vectorAdd, launch_config,
    d_A.get(), d_B.get(), d_C.get(), numElements
);    
cuda::memory::copy(h_C.get(), d_C.get(), size);

// (... verify results here...)

再次 - 通过抛出异常检查和报告所有潜在错误.这段代码用我的

用于CUDA运行时API库(Github)的精简Modern-C++包装器

请注意,异常在失败的调用之后带有字符串说明和CUDA运行时API状态代码.

使用这些包装器自动检查CUDA错误的一些链接:

• 一个测试程序抛出并捕获一堆异常
• 与错误相关的功能的文档

Answer 4

这里讨论的解决方案对我来说效果很好.该解决方案使用内置的cuda功能,实现起来非常简单.

相关代码复制如下:

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

__global__ void foo(int *ptr)
{
  *ptr = 7;
}

int main(void)
{
  foo<<<1,1>>>(0);

  // make the host block until the device is finished with foo
  cudaDeviceSynchronize();

  // check for error
  cudaError_t error = cudaGetLastError();
  if(error != cudaSuccess)
  {
    // print the CUDA error message and exit
    printf("CUDA error: %s\n", cudaGetErrorString(error));
    exit(-1);
  }

  return 0;
}