CUDA approfondi

Aucune fonction de CUDA ne renvoie directement ses erreurs. Il faut faire appel à une fonction spéciale, cudaGetLastError(), qui renvoie le code d'erreur, sous forme de cudaError_t.

Toutes les fonctions du runtime renvoient donc un code d'erreur, mais dans le cas d'appels asynchrones, comme le kernel retourne immédiatement, on ne peut pas savoir si une erreur a eu lieu. Vous devez donc utiliser une fonction de synchronisation afin de pouvoir récupérer l'erreur.

Chaque erreur est effacée par la suivante et une réussite renvoie aussi un code d'erreur. Récupérer une erreur remet le compteur à la réussite.

Un kernel ne peut pas retourner de code d'erreur, on doit donc vérifier s'il peut s'exécuter sans problème avant de le lancer.

Il existe une seconde fonction pour gérer les erreurs : cudaGetErrorString(), qui prend en paramètre un code d'erreur, tel que retourné par cudaGetLastError(). Elle retourne, sous la forme d'un const char *, la description de l'erreur.

cudaGetDeviceCount() et cudaGetDeviceProperties() permettent d'énumérer les périphériques, les cartes graphiques et leurs capacités.

int deviceCount;
cudaGetDeviceCount(& deviceCount);
int device;

for (device = 0 ; device < deviceCount ; ++device)
{
    cudaDeviceProp deviceProp;
    cudaGetDeviceProperties(&deviceProp, device);
}

Seuls les périphériques avec un indice de capacité de calcul supérieur ou égal à 1.0 sont retournés, les autres ne pouvant pas être exploités par CUDA.

cudaSetDevice() sert à sélectionner le périphérique qui exécutera les calculs du thread CPU.

cudaSetDevice(device);

Un périphérique doit être choisi avant d'appeler le moindre kernel, soit explicitement avec cudaSetDevice(), soit implicitement (le périphérique 0 est alors choisi). Tout appel ultérieur à cudaSetDevice() échouera.

La fonction cudaGetDeviceProperties() renvoie une structure ainsi définie.

struct cudaDeviceProp
{
    char name[256];
    size_t totalGlobalMem;
    size_t sharedMemPerBlock;
    int regsPerBlock;
    int warpSize;
    size_t memPitch;
    int maxThreadsPerBlock;
    int maxThreadsDim[3];
    int maxGridSize[3];
    size_t totalConstMem;
    int major;
    int minor;
    int clockRate;
    size_t textureAlignment;
    int deviceOverlap;
    int multiProcessorCount;
    int kernelExecTimeoutEnabled;
}

• name : chaîne ASCII identifiant le périphérique ;
• totalGlobalMem : mémoire globale totale disponible sur le périphérique en bytes ;
• sharedMemPerBlock : mémoire partagée disponible pour un bloc en bytes (cette mémoire est partagée par tous les blocs de threads résidant simultanément sur un multiprocesseur) ;
• regsPerBlock : nombre de registres 32 bits disponibles à un bloc de threads ;
• warpSize : taille d'un warp en threads ;
• memPitch : pas maximal en bytes permis par la copie de mémoire qui implique des régions de mémoire allouées par cudaMallocPitch(); ;
• maxThreadsPerBlock : nombre maximal de threads par bloc ;
• maxThreadsDim[3] : dimensions maximales d'un bloc ;
• maxGridSize[3] : dimensions maximales d'une grille ;
• totalConstMem : total de mémoire constante disponible sur le périphérique ;
• major : numéro de révision majeur de la capacité de calcul ;
• minor : numéro de révision mineur de la capacité de calcul ;
• clockRate : fréquence de l'horloge en kilohertz ;
• textureAlignment : alignement des textures requis, les adresses de base des textures qui sont alignées sur ceci n'ont pas besoin d'appliquer un offset lors des recherches ;
• deviceOverlap : 1 si le périphérique peut copier de la mémoire entre l'hôte et le périphérique en même temps qu'exécuter un kernel, 0 sinon ;
• multiProcessorCount : nombre de multiprocesseurs sur le périphérique ;
• kernelExecTimeoutEnabled : 1 s'il y a une limite de temps pour les kernels, 0 sinon.

Vous pouvez définir vous-même une structure de ce genre et demander à CUDA de choisir le périphérique qui y correspond le mieux, grâce à la fonction cudaChooseDevice( int * dev, const struct cudaDeviceProp * prop ) (elle retourne le périphérique dans dev).