Cuda hello world - Historique des versions

Jc.forton : /* Gestion unifiée */

2021-06-23T07:09:09Z

Gestion unifiée

← Version précédente		Version du 23 juin 2021 à 09:09
Ligne 162 :		Ligne 162 :
	</pre>		</pre>
	== Gestion unifiée ==		== Gestion unifiée ==
	[https://fr.wikipedia.org/wiki/Maxwell_(architecture_de_carte_graphique) L'architecture Maxwell] a introduit le support de la mémoire unifiée. Les variables allouées grâce à la fonction ''cudaMallocManaged'' voit leurs espaces mémoire transiter entre CPU et du GPU au besoin. Cella permet d'écrire les programmes plus simplement, sans avoir à appeler explicitement la fonction cudaMemcpy.		[https://fr.wikipedia.org/wiki/Maxwell_(architecture_de_carte_graphique) L'architecture Maxwell] a introduit le support de la mémoire unifiée. Les variables allouées grâce à la fonction ''cudaMallocManaged'' voit leurs espaces mémoire transiter entre CPU et GPU au besoin. Cella permet d'écrire les programmes plus simplement, sans avoir à appeler explicitement la fonction cudaMemcpy.

	Dans l'exemple suivant on créer une chaîne de caractère que l'on modifie dans le kernel GPU avant de l'afficher avec le CPU:		Dans l'exemple suivant on créer une chaîne de caractère que l'on modifie dans le kernel GPU avant de l'afficher avec le CPU:

Jc.forton : /* Manipulation mémoire */

2021-06-19T17:06:40Z

Manipulation mémoire

← Version précédente		Version du 19 juin 2021 à 19:06
Ligne 82 :		Ligne 82 :

	= Manipulation mémoire =		= Manipulation mémoire =
			== Gestion manuellement ==
	Le programme ci-dessous permet d'afficher le nom passé en paramètre :		Le programme ci-dessous permet d'afficher le nom passé en paramètre :
	<source lang="c">		<source lang="c">
Ligne 117 :		Ligne 118 :
	# on procède à la copie avec ''cudaMemCpy''		# on procède à la copie avec ''cudaMemCpy''
	# on procède à la libération mémoire avec ''cudaFree''		# on procède à la libération mémoire avec ''cudaFree''
	==allocation==		===allocation===
	''cudaMalloc'' prend en paramètre un pointeur côté CPU et la taille de l'allocation :		''cudaMalloc'' prend en paramètre un pointeur côté CPU et la taille de l'allocation :
	<source lang="c">		<source lang="c">
Ligne 123 :		Ligne 124 :
	</source>		</source>
	Le pointeur ''ptr'' permet de faire le lien entre les variables dans la mémoire centrale CPU et celle dans la mémoire graphique.		Le pointeur ''ptr'' permet de faire le lien entre les variables dans la mémoire centrale CPU et celle dans la mémoire graphique.
	==copie==		===copie===
	''cudaMemCpy'' prend en paramètre :		''cudaMemCpy'' prend en paramètre :
	* un pointeur source		* un pointeur source
Ligne 135 :		Ligne 136 :
	* cudaMemcpyHostToDevice (CPU → GPU)		* cudaMemcpyHostToDevice (CPU → GPU)
	* cudaMemcpyDeviceToHost (GPU → CPU)		* cudaMemcpyDeviceToHost (GPU → CPU)
	==libération==		===libération===
	''cudaFree'' prend en paramètre le pointeur côté GPU à libérer		''cudaFree'' prend en paramètre le pointeur côté GPU à libérer
	<source lang="c">		<source lang="c">
Ligne 159 :		Ligne 160 :
	CPU says : tala		CPU says : tala
	GPU says : tala		GPU says : tala
			</pre>
			== Gestion unifiée ==
			[https://fr.wikipedia.org/wiki/Maxwell_(architecture_de_carte_graphique) L'architecture Maxwell] a introduit le support de la mémoire unifiée. Les variables allouées grâce à la fonction ''cudaMallocManaged'' voit leurs espaces mémoire transiter entre CPU et du GPU au besoin. Cella permet d'écrire les programmes plus simplement, sans avoir à appeler explicitement la fonction cudaMemcpy.

			Dans l'exemple suivant on créer une chaîne de caractère que l'on modifie dans le kernel GPU avant de l'afficher avec le CPU:
			<source lang="c">
			#include <stdio.h>
			#include <stdlib.h>
			#include <cuda_runtime.h>
			#include <cuda.h>

			__global__ void say_hello(char *str){
			printf("CPU said : %s\n", str);
			memcpy(str, "world", 5);
			}

			int main(int argc, char * argv[]) {
			char *str;
			cudaMallocManaged(&str, 5 * sizeof(char));
			say_hello<<<1,1>>>(str);
			cudaDeviceSynchronize();
			printf("GPU said : %s\n", str);
			return EXIT_SUCCESS;
			}
			</source>
			La sortie permet de vérifier que l'espace mémoire est géré correctement:
			<pre>
			CPU said : hello
			GPU said : world
	</pre>		</pre>

Jc.forton : /* Manipulation mémoire */

2021-06-19T16:47:36Z

Manipulation mémoire

← Version précédente		Version du 19 juin 2021 à 18:47
Ligne 160 :		Ligne 160 :
	GPU says : tala		GPU says : tala
	</pre>		</pre>

	= Affichage des erreurs =		= Affichage des erreurs =
	L'exemple précédent met en évidence le manque de visibilité sur ce qui se passe côté GPU. Pour cela on peut utiliser la fonction ''gpuAssert'' pour afficher les erreurs lors de l’exécution du code sur le GPU:		L'exemple précédent met en évidence le manque de visibilité sur ce qui se passe côté GPU. Pour cela on peut utiliser la fonction ''gpuAssert'' pour afficher les erreurs lors de l’exécution du code sur le GPU:

Jc.forton : /* Parallélisation */

2021-06-15T07:00:44Z

Parallélisation

← Version précédente		Version du 15 juin 2021 à 09:00
Ligne 211 :		Ligne 211 :
	* blockIdx : id du bloc		* blockIdx : id du bloc
	* blockDim : dimension de chaque bloc		* blockDim : dimension de chaque bloc
			* gridDim : dimension de la grille (nombre de blocs)

	On peut schématiser la parallélisation effectuée par le GPU comme la matrice suivante:		On peut schématiser la parallélisation effectuée par le GPU comme la matrice suivante:
Ligne 224 :		Ligne 225 :

	__global__ void whoami() {		__global__ void whoami() {
	printf("blockDim [%d], blockIdx [%d], threadIdx[%d]\n", blockDim.x, blockIdx.x, threadIdx.x);		printf("gridDim[%d], blockDim [%d], blockIdx [%d], threadIdx[%d]\n", gridDim.x, blockDim.x, blockIdx.x, threadIdx.x);
	}		}

Ligne 235 :		Ligne 236 :
	Cette exemple donne la sortie suivante:		Cette exemple donne la sortie suivante:
	<pre>		<pre>
	blockDim [3], blockIdx [1], threadIdx[0]		gridDim[2], blockDim [3], blockIdx [0], threadIdx[0]
	blockDim [3], blockIdx [1], threadIdx[1]		gridDim[2], blockDim [3], blockIdx [0], threadIdx[1]
	blockDim [3], blockIdx [1], threadIdx[2]		gridDim[2], blockDim [3], blockIdx [0], threadIdx[2]
	blockDim [3], blockIdx [0], threadIdx[0]		gridDim[2], blockDim [3], blockIdx [1], threadIdx[0]
	blockDim [3], blockIdx [0], threadIdx[1]		gridDim[2], blockDim [3], blockIdx [1], threadIdx[1]
	blockDim [3], blockIdx [0], threadIdx[2]		gridDim[2], blockDim [3], blockIdx [1], threadIdx[2]
	</pre>		</pre>
	On peut remarquer que:		On peut remarquer que:
			* gridDim.x correspond au nombre de blocs (premier paramètre de la fonction chevron)
	* blockDim.x correspond au nombre de threads par blocs (deuxième paramètre de la fonction chevron)		* blockDim.x correspond au nombre de threads par blocs (deuxième paramètre de la fonction chevron)
	* blockIdx.x correspond à l'index du bloc (premier paramètre de la fonction chevron)		* blockIdx.x correspond à l'index du bloc (premier paramètre de la fonction chevron)
Ligne 251 :		Ligne 253 :
	__global__ void whoami() {		__global__ void whoami() {
	int index = blockDim.x * blockIdx.x + threadIdx.x;		int index = blockDim.x * blockIdx.x + threadIdx.x;
	printf("I am %d\n", index);		int total = gridDim.x * blockDim.x;
			printf("I am %d over %d\n", index, total);
	}		}
	</source>		</source>
	Ce qui donne le résultat suivant:		Ce qui donne le résultat suivant:
	<pre>		<pre>
	I am 3		I am 0 over 6
	I am 4		I am 1 over 6
	I am 5		I am 2 over 6
	I am 0		I am 3 over 6
	I am 1		I am 4 over 6
	I am 2		I am 5 over 6
	</pre>		</pre>

Jc.forton : /* Affichage des erreurs */

2021-06-04T20:43:55Z

Affichage des erreurs

@@ Ligne 161 : / Ligne 161 : @@
 </pre>
 = Affichage des erreurs =
-L'exemple suivant met en évidence le manque de visibilité sur ce qui se passe dans le GPU. Pour cela on peut utiliser les TODO
+L'exemple précédent met en évidence le manque de visibilité sur ce qui se passe côté GPU. Pour cela on peut utiliser la fonction ''gpuAssert'' pour afficher les erreurs lors de l’exécution du code sur le GPU:
 = Parallélisation =

Jc.forton : /* Manipulation mémoire */

2021-06-04T17:09:45Z

Manipulation mémoire

← Version précédente		Version du 4 juin 2021 à 19:09
Ligne 117 :		Ligne 117 :
	# on procède à la copie avec ''cudaMemCpy''		# on procède à la copie avec ''cudaMemCpy''
	# on procède à la libération mémoire avec ''cudaFree''		# on procède à la libération mémoire avec ''cudaFree''
			==allocation==
	''cudaMalloc'' prend en paramètre un pointeur côté CPU et la taille de l'allocation :		''cudaMalloc'' prend en paramètre un pointeur côté CPU et la taille de l'allocation :
	<source lang="c">		<source lang="c">
Ligne 123 :		Ligne 123 :
	</source>		</source>
	Le pointeur ''ptr'' permet de faire le lien entre les variables dans la mémoire centrale CPU et celle dans la mémoire graphique.		Le pointeur ''ptr'' permet de faire le lien entre les variables dans la mémoire centrale CPU et celle dans la mémoire graphique.
			==copie==
	''cudaMemCpy'' prend en paramètre :		''cudaMemCpy'' prend en paramètre :
	* un pointeur source		* un pointeur source
Ligne 135 :		Ligne 135 :
	* cudaMemcpyHostToDevice (CPU → GPU)		* cudaMemcpyHostToDevice (CPU → GPU)
	* cudaMemcpyDeviceToHost (GPU → CPU)		* cudaMemcpyDeviceToHost (GPU → CPU)
			==libération==

	''cudaFree'' prend en paramètre le pointeur côté GPU à libérer		''cudaFree'' prend en paramètre le pointeur côté GPU à libérer
	<source lang="c">		<source lang="c">
Ligne 161 :		Ligne 160 :
	GPU says : tala		GPU says : tala
	</pre>		</pre>
			= Affichage des erreurs =
			L'exemple suivant met en évidence le manque de visibilité sur ce qui se passe dans le GPU. Pour cela on peut utiliser les TODO

	= Parallélisation =		= Parallélisation =
	Le gros avantage du GPU c'est sa capacité à paralléliser l'exécution du code et pour cela on doit s'attarder sur la syntaxe chevron. On peut mixer le nombre de blocs ainsi que le nombre de thread par blocs jusqu'à un total de 1024, comme expliqué [https://forums.developer.nvidia.com/t/maximum-number-of-threads-on-thread-block/46392 ici] :		Le gros avantage du GPU c'est sa capacité à paralléliser l'exécution du code et pour cela on doit s'attarder sur la syntaxe chevron. On peut mixer le nombre de blocs ainsi que le nombre de thread par blocs jusqu'à un total de 1024, comme expliqué [https://forums.developer.nvidia.com/t/maximum-number-of-threads-on-thread-block/46392 ici] :

Jc.forton : /* Parallélisation */

2021-06-04T17:05:54Z

Parallélisation

@@ Ligne 162 : / Ligne 162 : @@
 </pre>
 = Parallélisation =
-Le gros avantage du GPU c'est sa capacité à paralléliser l'exécution du code et pour cela, il faut s'attarder sur la syntaxe chevron:
+Le gros avantage du GPU c'est sa capacité à paralléliser l'exécution du code et pour cela on doit s'attarder sur la syntaxe chevron. On peut mixer le nombre de blocs ainsi que le nombre de thread par blocs jusqu'à un total de 1024, comme expliqué [https://forums.developer.nvidia.com/t/maximum-number-of-threads-on-thread-block/46392 ici] :
 * on peut faire <<<1024, 1>>> ou <<<128,8>>> (128 * 8 = 1024)
 * on ne peut pas faire <<<256, 10>>> car 256 * 10 = 2560 > 1024
@@ Ligne 205 : / Ligne 202 : @@
 </pre>
 On peut remarquer que:
-* blockDim.x correspond au deuxième paramètre de la fonction chevron
+* blockDim.x correspond au nombre de threads par blocs (deuxième paramètre de la fonction chevron)
-* la valeur maximum prise par blockIdx.x correspond au premier paramètre de la fonction chevron (moins un car c'est l'index de la matrice)
+* blockIdx.x correspond à l'index du bloc (premier paramètre de la fonction chevron)
-* les valeurs de
+* threadIdx.x correspond à l'index du thread (compris entre 0 et blockDim.x - 1)

Jc.forton : /* Les bases */

2021-06-04T17:00:21Z

Les bases

← Version précédente		Version du 4 juin 2021 à 19:00
Ligne 19 :		Ligne 19 :
	L'API CUDA permet d'exécuter une fonction sur le GPU grâce à la notation chevrons ''<<<...>>>'' :		L'API CUDA permet d'exécuter une fonction sur le GPU grâce à la notation chevrons ''<<<...>>>'' :
	<source lang="c">		<source lang="c">
	fonction<<<~~CORE~~, ~~THREAD~~>>>();		fonction<<<NB_BLOCK, NB_THREAD>>>();
	</source>		</source>
	* fonction : correspond au nom de la fonction		* fonction : correspond au nom de la fonction
	* ~~CORE~~ : correspond au nombre de ~~cœur CUDA~~ utilisés		* NB_BLOCK : correspond au nombre de blocs utilisés
	* ~~THREAD~~ : correspond au nombre de threads par ~~cœur CUDA~~		* NB_THREAD : correspond au nombre de threads par bloc
	La fonction appelée doit être précédée du spécificateur ''__global__'' pour préciser au compilateur que sont exécution se fait sur le GPU:		La fonction appelée doit être précédée du spécificateur ''__global__'' pour préciser au compilateur que sont exécution se fait sur le GPU:
	<source lang="c">		<source lang="c">

Jc.forton : /* Parallélisation */

2021-06-04T16:59:33Z

Parallélisation

@@ Ligne 174 : / Ligne 174 : @@
 * blockDim  : dimension de chaque bloc
-Prenons l'exemple suivant:
+On peut schématiser la parallélisation effectuée par le GPU comme la matrice suivante:
 <source lang="c">
@@ Ligne 183 : / Ligne 186 : @@
 __global__ void whoami() {
-	int i = blockDim.x * blockIdx.x + threadIdx.x;
+	printf("blockDim [%d], blockIdx [%d], threadIdx[%d]\n", blockDim.x, blockIdx.x, threadIdx.x);
+}
@@ Ligne 195 : / Ligne 195 : @@
+}
 </source>

Jc.forton : /* Manipulation mémoire */

2021-06-04T16:22:18Z

Manipulation mémoire

@@ Ligne 109 : / Ligne 109 : @@
 CPU says : tala
 </pre>
-Il manque, encore une fois, l’exécution du code côté GPU... Ce qui est normale car la variable passée en paramètre est ''argv[1]'' qui est dans la mémoire centrale !
+Il manque, encore une fois, l’exécution du code côté GPU... Ce qui est normal car la variable passée en paramètre ''argv[1]'' est dans la mémoire centrale donc inaccessible côté GPU !
-Pour cela il faudrait copier le contenu de cette variable dans la mémoire du GPU. Cela se fait en deux étapes:
+Pour pouvoir l'utiliser, il faut copier le contenu de cette variable dans la mémoire du GPU.
 # on fait l'allocation mémoire avec ''cudaMalloc''
-# on procède à la copie avec ''cudaMemCpy
+# on procède à la copie avec ''cudaMemCpy''

Cuda hello world - Historique des versions

Jc.forton : /* Gestion unifiée */

Jc.forton : /* Manipulation mémoire */

Jc.forton : /* Manipulation mémoire */

Jc.forton : /* Parallélisation */

Jc.forton : /* Affichage des erreurs */

Jc.forton : /* Manipulation mémoire */

Jc.forton : /* Parallélisation */

Jc.forton : /* Les bases */

Jc.forton : /* Parallélisation */

Jc.forton : /* Manipulation mémoire */

Jc.forton : /* Gestion unifiée */

Jc.forton : /* Manipulation mémoire */

Jc.forton : /* Manipulation mémoire */

Jc.forton : /* Parallélisation */

Jc.forton : /* Affichage des erreurs */

Jc.forton : /* Manipulation mémoire */

Jc.forton : /* Parallélisation */

Jc.forton : /* Les bases */

Jc.forton : /* Parallélisation */

Jc.forton : /* Manipulation mémoire */