This shows you the differences between two versions of the page.
asc:laboratoare:06 [2024/04/10 12:02] emil.slusanschi [Executie asincrona Host si Device] |
asc:laboratoare:06 [2025/04/09 09:11] (current) alexandru.bala [Fluxuri nonimplicite] |
||
---|---|---|---|
Line 109: | Line 109: | ||
// init all elements to 0 | // init all elements to 0 | ||
- | cudaMemset(data, 0, NUM_ELEM); | + | cudaMemset(data, 0, NUM_ELEM * sizeof(unsigned long long int)); |
// launch kernel writes | // launch kernel writes | ||
Line 135: | Line 135: | ||
</code> | </code> | ||
- | Functie concurrentRW citeste valoarea de la adresa data[blockIdx.x], o incrementeaza cu threadId si apoi o scrie. | + | Functia concurrentRW citeste valoarea de la adresa data[blockIdx.x], o incrementeaza cu threadIdx.x si apoi o scrie. |
In acest caz avem 10 thread-uri care fac operatii citire/scriere la aceeasi adresa, deci un comportament nedefinit. | In acest caz avem 10 thread-uri care fac operatii citire/scriere la aceeasi adresa, deci un comportament nedefinit. | ||
Line 212: | Line 212: | ||
==== Executie asincrona Host si Device ==== | ==== Executie asincrona Host si Device ==== | ||
- | Folosind apeluri asincrone, operatiile de executie catre device sunt puse in coada avand controlul intors catre host instant. Astfel unitatea host poate continua executia fara sa fie blocata in asteptarea executiei. | + | Folosind apeluri asincrone, operatiile de executie catre device sunt puse in coada avand controlul intors catre host instant. Astfel unitatea host poate continua executia fara sa fie blocata in asteptarea altor task-uri. |
Urmatoarele operatii sunt asincrone relativ la host: | Urmatoarele operatii sunt asincrone relativ la host: | ||
- Lansari de kernel | - Lansari de kernel | ||
- | - Copieri in cadrul spatiului de memorie a unui device | + | - Copieri in cadrul spatiului de memorie al unui device |
- Copiere memorie host -> device, avand < 64 KB | - Copiere memorie host -> device, avand < 64 KB | ||
- Copiere memorie host -> device, avand functii cu sufix Async | - Copiere memorie host -> device, avand functii cu sufix Async | ||
Line 224: | Line 224: | ||
==== Fluxuri nonimplicite ==== | ==== Fluxuri nonimplicite ==== | ||
- | Pentru a folosi cudaMemcpyAsync, este necesar lucrul cu fluxuri nonimplictie (non-default streams), care, in C/C++ pot fi declarate, create si distruse in partea de cod de pe host (CPU) in urmatorul fel: | + | Pentru a folosi cudaMemcpyAsync, este necesar lucrul cu fluxuri nonimplicite (non-default streams), care, in C/C++ pot fi declarate, create si distruse in partea de cod de pe host (CPU) in urmatorul fel: |
<code C> | <code C> | ||
cudaStream_t stream1; | cudaStream_t stream1; | ||
cudaError_t result; | cudaError_t result; | ||
- | result = cudaStreamCreate(&stream1) | + | result = cudaStreamCreate(&stream1); |
- | result = cudaStreamDestroy(stream1) | + | result = cudaStreamDestroy(stream1); |
</code> | </code> | ||
Line 236: | Line 237: | ||
<code C> | <code C> | ||
- | result = cudaMemcpyAsync(d_a, a, N, cudaMemcpyHostToDevice, stream1) | + | // num_bytes = N * sizeof (type_a); |
+ | result = cudaMemcpyAsync(d_a, a, num_bytes, cudaMemcpyHostToDevice, stream1); | ||
</code> | </code> | ||
- | Pentru a emite un kernel către un flux nonimplicit, specificăm identificatorul fluxului ca al patrulea parametru de configurare a execuției. Se observă și un al treilea parametru de configurare a execuției, care este folosit pentru a aloca memorie partajată device-ului (GPU-ului), utilizându-se 0 dacă nu se dorește acest aspect. | + | Pentru a emite un kernel către un flux nonimplicit, specificăm identificatorul fluxului ca al patrulea parametru de configurare a execuției. Se observă și un al treilea parametru de configurare a execuției, care este folosit pentru a aloca memorie partajată (shared memory) device-ului (GPU-ului), utilizându-se 0 dacă nu se dorește acest aspect. |
<code C> | <code C> | ||
- | increment<<<1,N,0,stream1>>>(d_a) | + | increment<<<1,N,0,stream1>>>(d_a); |
</code> | </code> | ||