Каков канонический способ проверки на наличие ошибок с помощью API среды выполнения CUDA?

Просматривая ответы и комментарии на вопросы CUDA и вики-тег CUDA, я вижу, что часто предлагается проверять состояние возврата каждого вызова API на наличие ошибок. Документация API содержит такие функции, как cudaGetLastError, cudaPeekAtLastError, а также cudaGetErrorString, но каков наилучший способ собрать их вместе, чтобы надежно отлавливать и сообщать об ошибках, не требуя большого количества дополнительного кода?

Источник

5 ответов

Решение

Вероятно, лучший способ проверить наличие ошибок в коде API времени выполнения - определить функцию-обработчик стиля assert и макрос-обертку, например:

#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 Вызов может вернуть либо ошибки, которые произошли во время выполнения ядра, либо ошибки из самой копии памяти. Это может сбить с толку новичка, и я бы порекомендовал использовать явную синхронизацию после запуска ядра во время отладки, чтобы было легче понять, где могут возникнуть проблемы.

Источник

Ответ talonmies выше - отличный способ прервать приложение в assertстиль манеры

Иногда мы можем захотеть сообщить и исправить состояние ошибки в контексте C++ как часть более крупного приложения.

Вот довольно краткий способ сделать это, создав исключение C++, полученное из std::runtime_error с помощью thrust::system_error:

#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 это 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;
}
Источник

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...)

Опять же - все потенциальные ошибки проверяются и сообщаются через выброшенное исключение. Этот код использует мой

Оболочки Thin Modern-C++ для библиотеки CUDA Runtime API (Github)

Обратите внимание, что после неудачного вызова исключения содержат как строковое объяснение, так и код состояния API времени выполнения CUDA.

Несколько ссылок на то, как ошибки CUDA автоматически проверяются с помощью этих оболочек:

Источник

Обсуждаемое здесь решение хорошо сработало для меня. Это решение использует встроенные функции 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;
}
Источник

Позвольте мне добавить мой любимый макрос для проверки ошибок CUDA.

      #define CUDACHECK(err) do { cuda_check((err), __FILE__, __LINE__); } while(false)
inline void cuda_check(cudaError_t error_code, const char *file, int line)
{
    if (error_code != cudaSuccess)
    {
        fprintf(stderr, "CUDA Error %d: %s. In file '%s' on line %d\n", error_code, cudaGetErrorString(error_code), file, line);
        fflush(stderr);
        exit(error_code);
    }
}

Оберните каждый вызов функции CUDA вCUDACHECK()макрос и после каждого запуска ядра используйтеCUDACHECK(cudaPeekAtLastError()).

The do{...}while(false)в основном для того, чтобы поставить точку с запятой после макроса. Прочтите эту статью , чтобы получить отличное объяснение, почему она используется.

Источник
Другие вопросы по тегам