Cuda 16 12 09

  • Published on
    23-Jun-2015

  • View
    3.456

  • Download
    0

Embed Size (px)

DESCRIPTION

Lezione tenuta dal Dr. Francesco Rossi sulla programmazione scientifica usando le GPU

Transcript

<ul><li> 1. Programmazione di GPUs con CUDA:Un nuovo approccio al parallelismo Francesco Rossi mail.francesco.rossi@gmail.com 16/12/2009 Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 1 / 45 </li></ul><p> 2. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 2 / 45 3. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 3 / 45 4. I limiti sici delle architetture scalari Un processore ` un chip di silicio con dei transistor stampati.e Processori pi` potenti grazie alla miniaturizzazione dei transistor, diu feature size :Se 0 : Il numero dei transistor scala come N 12 Il tempo di calcolo solo come 1 E pi` facile costruire macchine capaci di compiere pi` e pi` task inuu u parallelo che uno pi` velocemente. u Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA:16/12/2009 4 / 45 5. I limiti sici delle architetture scalari Un processore ` un chip di silicio con dei transistor stampati.e Processori pi` potenti grazie alla miniaturizzazione dei transistor, diu feature size :Se 0 : Il numero dei transistor scala come N 12 Il tempo di calcolo solo come 1 E pi` facile costruire macchine capaci di compiere pi` e pi` task inuu u parallelo che uno pi` velocemente. u Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA:16/12/2009 4 / 45 6. Storia del parallelismo Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 5 / 45 7. Meglio una bassa latenza o un alto troughput? CPUs: Ottimizzate per avere accesso a bassa latenza ad estese memorie cache. Molti transistor impegnati nella predizione dei branches. Drawbacks: Parallelismo massivo inibito. Illusione della serialit`. a Scarsa potenza di calcolo. GPUs: Ottimizzate per computazioni parallele, pi` transistors dedicati alleu operazioni di calcolo. Latenze ammortizzate dallesecuzione concorrente di pi` threads u (SIMT). Ecienza derivante da capacit` di tenere locali i dati. a Alto troughput e parallelismo garantiti dalla presenza di centinaia di unit` di calcolo. aFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 6 / 45 8. Meglio una bassa latenza o un alto troughput? CPUs: Ottimizzate per avere accesso a bassa latenza ad estese memorie cache. Molti transistor impegnati nella predizione dei branches. Drawbacks: Parallelismo massivo inibito. Illusione della serialit`. a Scarsa potenza di calcolo. GPUs: Ottimizzate per computazioni parallele, pi` transistors dedicati alleu operazioni di calcolo. Latenze ammortizzate dallesecuzione concorrente di pi` threads u (SIMT). Ecienza derivante da capacit` di tenere locali i dati. a Alto troughput e parallelismo garantiti dalla presenza di centinaia di unit` di calcolo. aFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 6 / 45 9. Meglio una bassa latenza o un alto troughput? Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 7 / 45 10. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 8 / 45 11. Struttura di una GPUTerminologia comune: CPU (central processing unit) Host: il computer su cui ` installata la GPUe GPU (graphics processing unit) Device: scheda video/GPU. Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 9 / 45 12. Struttura di una GPUIl device ha una RAM (device memory) o-chip ma sulla scheda.Il processore (chip) ` formato da centinaia di processori scalari (SP). eGli SP sono raggruppati sicamente in gruppi da 8 in Multiprocessors.Gli SP in uno stesso multiprocessor condividono la stessa shared memory. Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 10 / 45 13. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 11 / 45 14. Threads Una GPU mantiene in esecuzione migliaia di threads perammortizzare le latenze dei singoli.Un thread ` unistanza di un algoritmo che viene fatta eseguire su unaeporzione dei dati. Lo stato di un thread ` denito da:e Il valore nei suoi registri. Il suo indice. La posizione dellistruzione che sta eseguendo ( instruction pointer). Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 12 / 45 15. Threads Una GPU mantiene in esecuzione migliaia di threads perammortizzare le latenze dei singoli.Un thread ` unistanza di un algoritmo che viene fatta eseguire su unaeporzione dei dati. Lo stato di un thread ` denito da:e Il valore nei suoi registri. Il suo indice. La posizione dellistruzione che sta eseguendo ( instruction pointer). Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 12 / 45 16. KernelsLinsieme dei threads che risolvono un problema ` chiamato kernel, oegriglia di threads. !Ogni thread allinterno del kernel (griglia) ` identicato da un indice. e Un kernel viene lanciato dallhost su uno o pi` device.u Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 13 / 45 17. BlocksI vari device eseguono i threads raggruppandoli in blocks (vicinati).I threads in un block condividono la shared memory.Un block ` lanalogo software del multiprocessor, ma un emultiprocessor ha allocati pi` blocks. u Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 14 / 45 18. In sintesi Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 15 / 45 19. In sintesi Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 16 / 45 20. In sintesi Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 17 / 45 21. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 18 / 45 22. CUDAFramework proprietario di NVIDIA per la programmazione di GPUNVIDIA.Pro: Ben documentato. Strutturato come estensione al linguaggio C. Compatibile con C++, templates. Multipiattaforma Esistono port per multicore. Molte librerie ottimizzate (cuFFT, cublas, cudpp, thrust). Standard de-facto.Contro: Lock-in? Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 19 / 45 23. OpenCL Framework standard open per la programmazione di GPU earchitetture parallele massive in genere.Pro: Funziona su tutte le schede e anche sulle CPU multicore. E uno standard open sviluppato da tutte le maggiori case produttrici/sviluppatrici.Contro: Ancora poche librerie rilasciate. Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 20 / 45 24. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 21 / 45 25. cudaMalloc!le memorie device e host sono RAM diverse e separate. Serve per allocare memoria sul device.Simile a malloc sullhost. oat *device pointer=0;cudaMalloc(&amp;device pointer, sizeof(oat)*N); serve per allocare sulla GPU un array di N oats. Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 22 / 45 26. cudaMalloc!le memorie device e host sono RAM diverse e separate. Serve per allocare memoria sul device.Simile a malloc sullhost. oat *device pointer=0;cudaMalloc(&amp;device pointer, sizeof(oat)*N); serve per allocare sulla GPU un array di N oats. Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 22 / 45 27. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 23 / 45 28. cudaMemcpy Sintassi: cudaMemcpy(puntatore a destinazione, puntatore sorgente,dimensione in bytes, ag cuda)Copia un segmento di memoria (nellesempio copia N oats): dallhost al device :cudaMemcpy(device ptr, host ptr, sizeof(oat)*N, cudaMemcpyHostToDevice); dal device allhost:cudaMemcpy(host ptr, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToHost); dal device al device o dallhost allhost (equivalente a memcpy):cudaMemcpy(device ptr2, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToDevice);Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 24 / 45 29. cudaMemcpy Sintassi: cudaMemcpy(puntatore a destinazione, puntatore sorgente,dimensione in bytes, ag cuda)Copia un segmento di memoria (nellesempio copia N oats): dallhost al device :cudaMemcpy(device ptr, host ptr, sizeof(oat)*N, cudaMemcpyHostToDevice); dal device allhost:cudaMemcpy(host ptr, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToHost); dal device al device o dallhost allhost (equivalente a memcpy):cudaMemcpy(device ptr2, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToDevice);Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 24 / 45 30. cudaMemcpy Sintassi: cudaMemcpy(puntatore a destinazione, puntatore sorgente,dimensione in bytes, ag cuda)Copia un segmento di memoria (nellesempio copia N oats): dallhost al device :cudaMemcpy(device ptr, host ptr, sizeof(oat)*N, cudaMemcpyHostToDevice); dal device allhost:cudaMemcpy(host ptr, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToHost); dal device al device o dallhost allhost (equivalente a memcpy):cudaMemcpy(device ptr2, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToDevice);Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 24 / 45 31. cudaMemcpy Sintassi: cudaMemcpy(puntatore a destinazione, puntatore sorgente,dimensione in bytes, ag cuda)Copia un segmento di memoria (nellesempio copia N oats): dallhost al device :cudaMemcpy(device ptr, host ptr, sizeof(oat)*N, cudaMemcpyHostToDevice); dal device allhost:cudaMemcpy(host ptr, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToHost); dal device al device o dallhost allhost (equivalente a memcpy):cudaMemcpy(device ptr2, device ptr, sizeof(oat)*N, cudaMemcpyDeviceToDevice);Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 24 / 45 32. cudaMemcpy Francesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 25 / 45 33. Outline1 Introduzione Perch` le GPUe 2 Programmare le GPU Struttura di una GPU Threads, kernels, blocks Due frameworks: CUDA e OpenCL 3 Primi passi con CUDA Allocare la memoria Sincronizzare i dati Lanciare lelaborazione Un esempio pi` praticou 4 ApplicazioniMetodi PIC per la simulazione di fenomeni laser-plasmaRaytracingFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 26 / 45 34. Ambiti delle funzioni !GPU e CPU eseguono codice separato:non ` possibile eseguire il codice CPU su una GPU.e Il codice per GPU e CPU pu` essere scritto nello stesso le .cu.o Ogni funzione C (in CUDA) ha un su ambito, a seconda del quale pu` oessere eseguita: sulla CPU pu` accedere solo alla memoria host o sulla GPU pu` accedere solo alla memoria device o A seconda dellambito di funzione, essa pu` essere chiamata:o dalla CPU dalla GPUFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 27 / 45 35. Ambiti delle funzioni !GPU e CPU eseguono codice separato:non ` possibile eseguire il codice CPU su una GPU.e Il codice per GPU e CPU pu` essere scritto nello stesso le .cu.o Ogni funzione C (in CUDA) ha un su ambito, a seconda del quale pu` oessere eseguita: sulla CPU pu` accedere solo alla memoria host o sulla GPU pu` accedere solo alla memoria device o A seconda dellambito di funzione, essa pu` essere chiamata:o dalla CPU dalla GPUFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 27 / 45 36. Ambiti delle funzioni !GPU e CPU eseguono codice separato:non ` possibile eseguire il codice CPU su una GPU.e Il codice per GPU e CPU pu` essere scritto nello stesso le .cu.o Ogni funzione C (in CUDA) ha un su ambito, a seconda del quale pu` oessere eseguita: sulla CPU pu` accedere solo alla memoria host o sulla GPU pu` accedere solo alla memoria device o A seconda dellambito di funzione, essa pu` essere chiamata:o dalla CPU dalla GPUFrancesco Rossimail.francesco.rossi@gmail.com ()Programmazione di GPUs con CUDA: 16/12/2009 27 / 45 37. Indicare lambito I qualicatori intrinseci del compilatore CUDA global deviceehost servono a specicare lambito di una funzione C. !Le funzioni che vengono eseguite sulla GPU sono quelle di ambitodevice e global . Dalla GPU si possono chiamare solo le funzioni dichiarate come device . Dalla CPU si possono chiamare: Le funzioni host (il qualicatore ` acceso di default se ` omesso ee...</p>