OpenACC Tutorial - Adding directives/fr: Difference between revisions

Updating to match new version of source page
(Created page with "[https://developer.nvidia.com/nvidia-visual-profiler NVIDIA Visual Profiler (NVVP)] est un profileur graphique pour les applications OpenACC. C'est un nutil d'analyse pour les...")
(Updating to match new version of source page)
 
(20 intermediate revisions by 3 users not shown)
Line 13: Line 13:


== Transfert vers un processeur graphique (GPU) ==  
== Transfert vers un processeur graphique (GPU) ==  
Avant de porter du code sur un GPU, il faut savoir qu'ils ne partagent pas la même mémoire que le CPU de l'hôte.
Avant de porter du code sur un GPU, il faut savoir que ceux-ci ne partagent pas la même mémoire que le CPU de l'hôte.
* la mémoire de l'hôte est en général plus grande, mais plus lente que la mémoire du GPU;
* la mémoire de l'hôte est en général plus grande, mais plus lente que la mémoire du GPU;
* un GPU n'a pas d'accès direct à la mémoire de l'hôte;
* un GPU n'a pas d'accès direct à la mémoire de l'hôte;
Line 66: Line 66:
Quand le compilateur lit la directive OpenACC <tt>kernels</tt>, il analyse le code pour identifier les sections pouvant être parallélisées.
Quand le compilateur lit la directive OpenACC <tt>kernels</tt>, il analyse le code pour identifier les sections pouvant être parallélisées.
Ceci correspond souvent au corps d'une boucle qui a des itérations indépendantes.
Ceci correspond souvent au corps d'une boucle qui a des itérations indépendantes.
Dans ce cas, le compilateur délimite le début et la fin du corps du code avec la fonction nommée [https://en.wikipedia.org/wiki/Compute_kernel ''kernel''].
Dans ce cas, le compilateur délimite le début et la fin du corps du code avec la fonction [https://en.wikipedia.org/wiki/Compute_kernel ''kernel''].
Les appels à cette fonction ne seront pas affectés par les autres appels.
Les appels à cette fonction ne seront pas affectés par les autres appels.
La fonction est compilée et peut ensuite être exécutée sur un accélérateur.
La fonction est compilée et peut ensuite être exécutée sur un accélérateur.
Line 114: Line 114:
|title=Description ou prescription
|title=Description ou prescription
|content=
|content=
Si vous avez déjà utilisé [[OpenMP/fr OpenMP]], vous retrouverez dans OpenACC le principe de ''directives''. Il existe cependant d'importantes différences entre les directives OpenMP et OpenACC&nbsp;:
Si vous avez déjà utilisé [[OpenMP/fr|OpenMP]], vous retrouverez dans OpenACC le principe de ''directives''. Il existe cependant d'importantes différences entre les directives OpenMP et OpenACC&nbsp;:
* Les directives OpenMP sont à la base ''prescriptives''. Ceci signifie que le compilateur est forcé d'accomplir la parallélisation, peu importe que l'effet détériore ou améliore la performance. Le résultat est prévisible pour tous les compilateurs. De plus, la parallélisation se fera de la même manière, peu importe le matériel utilisé pour exécuter le code. Par contre, le même code peut connaitre une moins bonne performance, dépendant de l'architecture. Il peut donc être préférable par exemple de changer l'ordre des boucles. Pour paralléliser du code avec OpenMP et obtenir une performance optimale dans différentes architectures, il faudrait avoir un ensemble différent de directives pour chaque architecture.  
* Les directives OpenMP sont à la base ''prescriptives''. Ceci signifie que le compilateur est forcé d'accomplir la parallélisation, peu importe que l'effet détériore ou améliore la performance. Le résultat est prévisible pour tous les compilateurs. De plus, la parallélisation se fera de la même manière, peu importe le matériel utilisé pour exécuter le code. Par contre, le même code peut connaitre une moins bonne performance, dépendant de l'architecture. Il peut donc être préférable par exemple de changer l'ordre des boucles. Pour paralléliser du code avec OpenMP et obtenir une performance optimale dans différentes architectures, il faudrait avoir un ensemble différent de directives pour chaque architecture.  


Line 160: Line 160:
==== Construire avec OpenACC ====
==== Construire avec OpenACC ====


Les compilateurs NVidia utilisent l'option <code>-ta</code> (''target accelerator'') pour permettre la compilation pour un accélérateur. Nous utilisons la sous-option <code>tesla:managed</code> pour indiquer au compilateur que nous voulons compiler pour des GPU Tesla et que nous voulons utiliser la [https://developer.nvidia.com/blog/unified-memory-cuda-beginners/ mémoire gérée]  pour simplifier le transfert de données en provenance et à destination du périphérique; nous n'utiliserons pas cette option dans un prochain exemple. Nous utilisons aussi l'option <code>-fast</code> pour l'optimisation.
Les compilateurs NVidia utilisent l'option <code>-acc</code> pour permettre la compilation pour un accélérateur. Nous utilisons la sous-option <code>-gpu=managed</code> pour indiquer au compilateur que nous voulons utiliser la [https://developer.nvidia.com/blog/unified-memory-cuda-beginners/ mémoire gérée]  pour simplifier le transfert de données en provenance et à destination du périphérique; nous n'utiliserons pas cette option dans un prochain exemple. Nous utilisons aussi l'option <code>-fast</code> pour l'optimisation.


{{Command
{{Command
|nvc++ -fast -Minfo{{=}}accel -ta{{=}}tesla:managed main.cpp -o challenge
|nvc++ -fast -Minfo{{=}}accel -acc -gpu{{=}}managed main.cpp -o challenge
|result=
|result=
...
...
Line 198: Line 198:


=== Boucle avec clause <tt>independent</tt> ===  
=== Boucle avec clause <tt>independent</tt> ===  
Une autre façon de s'assurer que le compilateur traite les boucles de manière indépendante est de le spécifier explicitement  avec la clause <tt>independent</tt>. Comme toute autre directive ''prescriptive'', le compilateur y est obligé et l'analyse qu'il pourrait faire ne sera pas considérée. En reprenant l'exemple de la section ''La directive <tt>kernels</tt>'', nous avons&nbsp;:   
Une autre façon de s'assurer que le compilateur traite les boucles de manière indépendante est de le spécifier explicitement  avec la clause <tt>independent</tt>. Comme toute autre directive ''prescriptive'', le compilateur y est obligé et l'analyse qu'il pourrait faire ne sera pas considérée. En reprenant l'exemple de la section ''La directive <tt>kernels</tt>'' ci-dessus, nous avons&nbsp;:   
<syntaxhighlight lang="cpp" line highlight="3">
<syntaxhighlight lang="cpp" line highlight="3">
#pragma acc kernels
#pragma acc kernels
Line 225: Line 225:
Remarquez que les autres pointeurs n'ont pas besoin d'être restreints puisque le compilateur ne les rapporte pas comme causant des problèmes. En recompilant avec les changements que nous venons de faire, le compilateur émet le message suivant&nbsp;:  
Remarquez que les autres pointeurs n'ont pas besoin d'être restreints puisque le compilateur ne les rapporte pas comme causant des problèmes. En recompilant avec les changements que nous venons de faire, le compilateur émet le message suivant&nbsp;:  
{{Command
{{Command
|nvc++ -fast -Minfo{{=}}accel -ta{{=}}tesla:managed main.cpp -o challenge
|nvc++ -fast -Minfo{{=}}accel -acc -gpu{{=}}managed main.cpp -o challenge
|result=
|result=
matvec(const matrix &, const vector &, const vector &):
matvec(const matrix &, const vector &, const vector &):
Line 282: Line 282:


[https://developer.nvidia.com/nvidia-visual-profiler NVIDIA Visual Profiler (NVVP)] est un profileur graphique pour les applications OpenACC.
[https://developer.nvidia.com/nvidia-visual-profiler NVIDIA Visual Profiler (NVVP)] est un profileur graphique pour les applications OpenACC.
C'est un nutil d'analyse pour les '''codes écrits avec les directives OpenACC et CUDA C/C++'''.
C'est un outil d'analyse pour les '''codes écrits avec les directives OpenACC et CUDA C/C++'''.
En conséquence, si l'exécutable n'utilise pas le GPU, ce profileur ne fournira aucun résultat.
En conséquence, si l'exécutable n'utilise pas le GPU, ce profileur ne fournira aucun résultat.


When [[Visualization/en#Remote_windows_with_X11-forwarding|X11 is forwarded to an X-Server]], or when using a [[VNC|Linux desktop environment]] (also via [[JupyterHub#Desktop|JupyterHub]] with two (2) CPU cores, 5000M of memory and one (1) GPU),
Quand [[Visualization#Fenêtres_à_distance_avec_redirection_X11|X11 est redirigé vers un serveur X-Server]] ou quand vous utilisez un [[VNC/fr|environnement bureau Linux]] (aussi via [[JupyterHub/fr#Bureau|JupyterHub]] avec 2 cœurs CPU, 5000M de mémoire et 1 GPU),
it is possible to launch the NVVP from a terminal:
vous pouvez lancer NVVP à partir d'un terminal&nbsp;:
{{Command
{{Command
|module load cuda/11.7 java/1.8
|module load cuda/11.7 java/1.8
Line 294: Line 294:
}}
}}


# After the NVVP startup window, you get prompted for a ''Workspace'' directory, which will be used for temporary files. Replace <code>home</code> with <code>scratch</code> in the suggested path. Then click ''OK''.
# Après l'affichage de la fenêtre de lancement de NVVP, vous devez entrer le répertoire ''Workspace'' qui sera employé pour les fichiers temporaires. Dans le chemin suggéré, remplacez <code>home</code> par <code>scratch</code> et cliquez sur ''OK''.
# Select ''File > New Session'', or click on the corresponding button in the toolbar.
# Sélectionnez ''File > New Session'' ou cliquez sur le bouton correspondant dans la barre d'outils.
# Click on the ''Browse'' button at the right of the ''File'' path editor.
# Cliquez sur le bouton ''Browse'' à la droite du champ ''File'' pour le chemin.
## Change directory if needed.
## Changez le répertoire s'il y a lieu.
## Select an executable built from codes written with OpenACC and CUDA C/C++ instructions.
## Sélectionnez un exécutable construit avec des codes écrits avec des directives OpenACC et CUDA C/C++.
# Below the ''Arguments'' editor, select the profiling option ''Profile current process only''.
# Sous le champ ''Arguments'', sélectionnez l'option ''Profile current process only''.
# Click ''Next >'' to review additional profiling options.
# Cliquez sur ''Next >'' pour voir les autres options de profilage.
# Click ''Finish'' to start profiling the executable.
# Cliquez sur ''Finish'' pour lancer le profilage de l'exécutable.


This can be done with the following steps:  
Pour faire ceci, suivez ces étapes&nbsp;:  
# Start <tt>nvvp</tt> with the command <tt>nvvp &</tt>  (the <tt>&</tt> sign is to start it in the background)
# Lancez <tt>nvvp</tt> avec la commande <tt>nvvp &</tt>  (le symbole <tt>&</tt> commande le lancement en arrière-plan).
# Go in File -> New Session
# Sélectionnez '' File -> New Session''.
# In the "File:" field, search for the executable (named <tt>challenge</tt> in our example).
# Dans le champ ''File:'', cherchez l'exécutable (nommé dans notre exemple <tt>challenge</tt>).
# Click "Next" until you can click "Finish".  
# Cliquez sur ''Next'' jusqu'à ce que vous puissiez cliquer sur ''Finish''.  


Le programme est exécuté et on obtient un tableau chronologique du déroulement (voir l'image). On remarque que le transfert de données entre le départ et l'arrivée occupe la plus grande partie du temps d'exécution, ce qui est fréquent quand du code est porté d'un CPU vers un GPU. Nous verrons comment ceci peut être amélioré dans la prochaine partie, [https://docs.computecanada.ca/wiki/OpenACC_Tutorial_-_Data_movement/fr Mouvement des données].
Le programme est exécuté et on obtient un tableau chronologique du déroulement (voir l'image). On remarque que le transfert de données entre le départ et l'arrivée occupe la plus grande partie du temps d'exécution, ce qui est fréquent quand du code est porté d'un CPU vers un GPU. Nous verrons comment ceci peut être amélioré dans la prochaine partie, [[OpenACC Tutorial - Data movement/fr|Mouvement des données]].


== La directive <tt>parallel loop</tt> ==  
== La directive <tt>parallel loop</tt> ==  
Line 346: Line 346:
La compilation produit le message suivant&nbsp;:
La compilation produit le message suivant&nbsp;:
{{Command
{{Command
|pgc++ -fast -Minfo{{=}}accel -ta{{=}}tesla:managed main.cpp -o challenge
|nvc++ -fast -Minfo{{=}}accel -acc -gpu{{=}}managed main.cpp -o challenge
|result=
|result=
matvec(const matrix &, const vector &, const vector &):
matvec(const matrix &, const vector &, const vector &):
Line 359: Line 359:
           34, Loop is parallelizable
           34, Loop is parallelizable
}}
}}
==Différences entre <tt>parallel loop</tt> et <tt>kernels</tt>==
==Différences entre <tt>parallel loop</tt> et <tt>kernels</tt>==
{| class="wikitable" width="100%"
{| class="wikitable" width="100%"
Line 381: Line 382:
|content=
|content=
# Modifiez les fonctions <tt>matvec</tt>, <tt>waxpby</tt> et <tt>dot</tt>. Vous pouvez utiliser soit <tt>kernels</tt>, soit <tt>parallel loop</tt>. La solution se trouve dans les répertoires <tt>step1. de [https://github.com/calculquebec/cq-formation-openacc Github] *</tt>.  
# Modifiez les fonctions <tt>matvec</tt>, <tt>waxpby</tt> et <tt>dot</tt>. Vous pouvez utiliser soit <tt>kernels</tt>, soit <tt>parallel loop</tt>. La solution se trouve dans les répertoires <tt>step1. de [https://github.com/calculquebec/cq-formation-openacc Github] *</tt>.  
# Modifiez le Makefile en ajoutant <tt>-ta{{=}}tesla:managed</tt> et <tt>-Minfo{{=}}accel</tt> aux indicateurs pour le compilateur.  
# Modifiez le Makefile en ajoutant <tt>-acc -gpu{{=}}managed</tt> et <tt>-Minfo{{=}}accel</tt> aux indicateurs pour le compilateur.  
}}
}}


<div class="mw-translate-fuzzy">
[[OpenACC Tutorial - Profiling/fr|<- Page précédente, Profileurs]] | [[OpenACC Tutorial/fr|^- Retour au début du tutoriel]] | [[OpenACC Tutorial - Data movement/fr|Page suivante, Mouvement des données ->]]
[https://docs.computecanada.ca/wiki/OpenACC_Tutorial_-_Data_movement/fr Page suivante, Mouvement des données]<br>
[https://docs.computecanada.ca/wiki/OpenACC_Tutorial/fr Retour au début du tutoriel]
</div>
38,760

edits