Quelle est la manière canonique de vérifier les erreurs en utilisant l'API d'exécution de CUDA?

Question

En regardant à travers les réponses et les commentaires sur les questions CUDA, et dans le wiki tag CUDA , je vois qu'il est souvent suggéré que le statut de retour de chaque appel d'API doit être vérifié pour les erreurs. La documentation de l'API contient des fonctions telles que cudaGetLastError , cudaPeekAtLastError et cudaGetErrorString , mais quel est le meilleur moyen de rassembler ces erreurs de façon fiable et sans erreurs? de code supplémentaire?

Answer 1

Probablement la meilleure façon de vérifier les erreurs lors de l'exécution code de l'API est de définir une assertion de style fonction de gestionnaire et de wrapper macro comme ceci:

#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);
   }
}

Vous pouvez ensuite envelopper chaque appel d'API avec l' gpuErrchk macro, qui va traiter l'état de retour de l'appel d'API qu'il l'enveloppe, par exemple:

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

Si il y a une erreur dans un appel, un message texte décrivant l'erreur et le fichier et la ligne dans votre code où s'est produite l'erreur sera émise pour stderr et l'application se fermera. Vous pourrait éventuellement modifier gpuAssert de soulever une exception plutôt que d'appeler exit() plus sophistiqué demande si c'était nécessaire.

Une deuxième question connexe est de savoir comment vérifier les erreurs dans le noyau lance, qui ne peuvent pas être directement enveloppé dans un appel de macro comme norme d'exécution des appels d'API. Pour les noyaux, quelque chose comme ceci:

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

va d'abord vérifier invalide lancement argument, puis forcer l'hôte d'attendre jusqu'à ce que le noyau s'arrête et vérifie une erreur d'exécution. La synchronisation peut être éliminé si vous avez un blocage d'appel d'API comme ceci:

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

auquel cas l' cudaMemcpy appel peut renvoyer soit à des erreurs survenues lors de l'exécution du noyau ou de la copie de la mémoire elle-même. Cela peut être déroutant pour le débutant, et je recommande l'utilisation explicite de synchronisation après un noyau de lancement au cours du débogage pour le rendre plus facile à comprendre d'où des problèmes peuvent être soulevées.

Answer 2

talonmies' réponse ci-dessus est une excellente façon d'interrompre une application dans un assert-style.

Parfois, il peut être bon de signaler et de le récupérer à partir d'une condition d'erreur en C++ contexte dans le cadre d'une application plus large.

Voici une assez laconique façon de le faire qu'en lançant une des exceptions C++ dérivé d' std::runtime_error l'aide 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);
  }
}

Cela permettra d'intégrer le nom du fichier, le numéro de ligne, et en langue anglaise, la description de l' cudaError_t dans la levée d'une exception de l' .what() membre:

#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;
}

La sortie:

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

Un client de l' some_function peut distinguer CUDA erreurs des autres types d'erreurs si vous le souhaitez:

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;
}

Parce qu' thrust::system_error est std::runtime_error, on peut également gérer de la même manière d'une large classe d'erreurs si nous n'avons pas besoin de la précision de l'exemple précédent:

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

Pour ceux d'entre vous qui ne peuvent pas/ne pas utiliser les exceptions, voici une suggestion qui vous permet de faire des appels comme suit:

CUDA_CALL(cudaSomeFunction, param1, param2, param3);

En cas d'erreur, les messages d'erreur de mentionner le nom de l'appel d'API que vous avez utilisé; de plus, il sera gentil et bavard.

Voici le code:

#define CUDA_CALL(cuda_function, ...)  { \
    cudaError_t status = cuda_function(__VA_ARGS__); \
    cudaEnsureSuccess(status, #cuda_function, __FILE__, __LINE__); \
}

bool cudaEnsureSuccess(cudaError_t status, const char* status_context_description,
        bool die_on_error, bool debug_prints, char* filename, int line_number) {
    if (status_context_description == NULL)
        status_context_description = "";
    if (status == cudaSuccess) {
#if REPORT_CUDA_SUCCESS
         cerr <<  "Succeeded: " << status_context_description << std::endl << std::flush;
#endif
        return true;
    }
    const char* errorString = cudaGetErrorString(status);
    cerr << "CUDA Error: ";
    if (status_context_description != NULL) {
        cerr << status_context_description << ": ";
    }
    if (errorString != NULL) {
        cerr << errorString;
    }
    else {
        cerr << "(Unknown CUDA status code " << status << ")";
    }
    if ((filename != NULL) && (line_number != NULL)) {

        cerr << filename << ":" << line_number;
    }

    cerr << std::endl << std::flush;
    if(die_on_error) {
        exit(EXIT_FAILURE);
            // ... or cerr << "FATAL ERROR" << etc. etc.
    }
    return false;
}

Notes:

• La macro est nécessaire, depuis une simple fonction ne peut pas déterminer le CUDA appel d'API a été apportée sans l'aide de la réflexion fonctionnalités manquantes à partir de C++.
• Cette fonction peut également être utilisée avec un noyau d'exécution wrapper macro qui assure le succès.