Alliance Doc

CUDA tutorial/fr: Difference between revisions

Page Discussion

CUDA tutorial/fr (view source)

Revision as of 20:03, 1 March 2018

1,969 bytes added ,  6 years ago
Updating to match new version of source page
No edit summary
(Updating to match new version of source page)
Line 94: Line 94:
** Identique à cudaMemcpy, mais transfère les données de manière asynchrone, ce qui signifie que l'exécution des autres processus n'est pas bloquée.
** Identique à cudaMemcpy, mais transfère les données de manière asynchrone, ce qui signifie que l'exécution des autres processus n'est pas bloquée.


<div class="mw-translate-fuzzy">
= Exemple d’un programme CUDA C simple=
= Exemple d’un programme CUDA C simple=
Dans cet exemple, nous ajoutons deux nombres au GPU. Il s'agit d'un exemple très simple et il ne faut pas s'attendre à observer une grande accélération.
Dans cet exemple, nous ajoutons deux nombres au GPU. Il s'agit d'un exemple très simple et il ne faut pas s'attendre à observer une grande accélération.
<syntaxhighlight lang="cpp" line highlight="1,5">
<syntaxhighlight lang="cpp" line highlight="1,5">
__global__  void add (int *a, int *b, int *c){
__global__  void add (int *a, int *b, int *c){
</div>
<syntaxhighlight lang="cpp" line highlight="1,5">
___global__ void add (int *a, int *b, int *c){
  *c = *a + *b;
}


*c = *a + *b;
}
int main(void){
int main(void){
int a, b, c;
  int a, b, c;
int *dev_a, *dev_b, *dev_c;
  int *dev_a, *dev_b, *dev_c;
int size = sizeof(int);
  int size = sizeof(int);


//  allocate device copies of a,b, c
  //  allocate device copies of a,b, c
cudaMalloc ( (void**) &dev_a, size);
  cudaMalloc ( (void**) &dev_a, size);
cudaMalloc ( (void**) &dev_b, size);
  cudaMalloc ( (void**) &dev_b, size);
cudaMalloc ( (void**) &dev_c, size);
  cudaMalloc ( (void**) &dev_c, size);


a=2; b=7;
  a=2; b=7;
//  copy inputs to device
  //  copy inputs to device
cudaMemcpy (dev_a, &a, size, cudaMemcpyHostToDevice);
  cudaMemcpy (dev_a, &a, size, cudaMemcpyHostToDevice);
cudaMemcpy (dev_b, &b, size, cudaMemcpyHostToDevice);
  cudaMemcpy (dev_b, &b, size, cudaMemcpyHostToDevice);


// launch add() kernel on GPU, passing parameters
  // launch add() kernel on GPU, passing parameters
add <<< 1, 1 >>> (dev_a, dev_b, dev_c);
  add <<< 1, 1 >>> (dev_a, dev_b, dev_c);


// copy device result back to host
  // copy device result back to host
cudaMemcpy (&c, dev_c, size, cudaMemcpyDeviceToHost);
  cudaMemcpy (&c, dev_c, size, cudaMemcpyDeviceToHost);


cudaFree ( dev_a ); cudaFree ( dev_b ); cudaFree ( dev_c );  
  cudaFree ( dev_a ); cudaFree ( dev_b ); cudaFree ( dev_c );
}
}
</syntaxhighlight>
</syntaxhighlight>


<div class="mw-translate-fuzzy">
Il manque certainement quelque chose; ce code n’a pas une allure parallèle…  Comme solution, modifions le contenu du ''kernel'' entre les signes <<< et >>>.
Il manque certainement quelque chose; ce code n’a pas une allure parallèle…  Comme solution, modifions le contenu du ''kernel'' entre les signes <<< et >>>.


Line 134: Line 139:
<syntaxhighlight lang="cpp" line highlight="1,5">
<syntaxhighlight lang="cpp" line highlight="1,5">
__global__  void add (int *a, int *b, int *c){
__global__  void add (int *a, int *b, int *c){
 
</div>
c[blockIdx.x] = a[blockIdx.x] + b[blockIdx.x];
<syntaxhighlight lang="cpp" line highlight="1,5">
add <<< N, 1 >>> (dev_a, dev_b, dev_c);
</syntaxhighlight>
Here we replaced 1 by N, so that N different CUDA blocks will be executed at the same time. However, in order to achieve parallelism we need to make some changes to the kernel as well:
<syntaxhighlight lang="cpp" line highlight="1,5">
__global__  void add (int *a, int *b, int *c){
  c[blockIdx.x] = a[blockIdx.x] + b[blockIdx.x];
</syntaxhighlight>
</syntaxhighlight>
où blockIdx.x est le numéro unique identifiant un bloc CUDA. De cette manière, chaque bloc CUDA ajoute une valeur de a[ ] à b[ ].
where blockIdx.x is the unique number identifying a CUDA block. This way each CUDA block adds a value from a[ ] to b[ ].
[[File:Cuda-blocks-parallel.png|thumbnail|Parallélisation basée sur les blocs. ]]
[[File:Cuda-blocks-parallel.png|thumbnail|CUDA blocks-based parallelism. ]]


<div class="mw-translate-fuzzy">
Modifions à nouveau le contenu entre les crochets triples.  
Modifions à nouveau le contenu entre les crochets triples.  
<syntaxhighlight lang="cpp" line highlight="1,5">
<syntaxhighlight lang="cpp" line highlight="1,5">
Line 145: Line 157:
</syntaxhighlight>
</syntaxhighlight>
La tâche est maintenant distribuée sur des fils parallèles plutôt que sur des blocs. Quel est l'avantage des fils parallèles? Contrairement aux blocs, les fils peuvent communiquer ensemble; autrement dit, nous parallélisons sur plusieurs fils dans le bloc quand la communication est intense. Les portions de code qui peuvent être exécutées indépendamment, soit avec peu ou pas de communication, sont distribuées sur des blocs parallèles.
La tâche est maintenant distribuée sur des fils parallèles plutôt que sur des blocs. Quel est l'avantage des fils parallèles? Contrairement aux blocs, les fils peuvent communiquer ensemble; autrement dit, nous parallélisons sur plusieurs fils dans le bloc quand la communication est intense. Les portions de code qui peuvent être exécutées indépendamment, soit avec peu ou pas de communication, sont distribuées sur des blocs parallèles.
</div>
<syntaxhighlight lang="cpp" line highlight="1,5">
add <<< 1, '''N''' >>> (dev_a, dev_b, dev_c);
</syntaxhighlight>
Now instead of blocks, the job is distributed across parallel threads. What is the advantage of having parallel threads ? Unlike blocks, threads can communicate between each other: in other words, we parallelize across multiple threads in the block when heavy communication is involved. The chunks of code that can run independently, i.e. with little or no communication, are distributed across parallel blocks.


<div class="mw-translate-fuzzy">
= Avantages de la mémoire partagée=
= Avantages de la mémoire partagée=
Jusqu'ici, tous les transferts en mémoire dans le ''kernel'' ont été via la mémoire régulière (globale) du GPU, ce qui est relativement lent. Il y a souvent tellement de communication entre fils que la performance est significativement diminuée.  Pour contrer ce problème, nous pouvons utiliser la mémoire partagée qui peut accélérer les transferts en mémoire entre les fils. Le secret par contre est que seuls les fils du même bloc peuvent communiquer. Pour illustrer l'utilisation de cette mémoire, voyons l'exemple du produit scalaire où deux vecteurs sont multipliés élément par élément et additionnés par la suite, ainsi :
Jusqu'ici, tous les transferts en mémoire dans le ''kernel'' ont été via la mémoire régulière (globale) du GPU, ce qui est relativement lent. Il y a souvent tellement de communication entre fils que la performance est significativement diminuée.  Pour contrer ce problème, nous pouvons utiliser la mémoire partagée qui peut accélérer les transferts en mémoire entre les fils. Le secret par contre est que seuls les fils du même bloc peuvent communiquer. Pour illustrer l'utilisation de cette mémoire, voyons l'exemple du produit scalaire où deux vecteurs sont multipliés élément par élément et additionnés par la suite, ainsi :
Line 163: Line 181:
int sum; for(int i=0;i<N;i++) sum+= temp[i];
int sum; for(int i=0;i<N;i++) sum+= temp[i];
*c=sum; }  
*c=sum; }  
}
</syntaxhighlight>
</div>
<syntaxhighlight lang="cpp" line highlight="1,5">
__global__  void dot(int *a, int *b, int *c){
  int temp = a[threadIdx.x]*b[threadIdx.x];
}
</syntaxhighlight>
After each thread computes its portion, we need to add everything together: each thread has to share its data.  However, the problem is that each copy of thread's temp variable is private. This can be resolved by the use of shared memory. Below is the kernel with the modifications to use shared memory:
<syntaxhighlight lang="cpp" line highlight="1,4">
#define N 512
__global__  void dot(int *a, int *b, int *c){
  __shared__ int temp[N];
  temp[threadIdx.x] = a[threadIdx.x]*b[threadIdx.x];
  __syncthreads();
  if(threadIdx.x==0){
    int sum; for(int i=0;i<N;i++) sum+= temp[i];
    *c=sum;
  }
}
}
</syntaxhighlight>
</syntaxhighlight>
FuzzyBot
Bots
38,760

edits

Retrieved from "https://docs.alliancecan.ca/wiki/Special:MobileDiff/47648"