Programmeermodel voor algemene berekeningen op NVIDIA-GPU's. Kernels worden geschreven in een uitbreiding van C++ en parallellisme wordt uitgedrukt in de hiërarchie grid, block en thread.
Japanse benaming: CUDA
“CUDA” komt voor in 18 oefenvragen uit 1 onderwerpen, bijv. GPU-programmeren.
Vakgebied: GPU-programmeren
In CUDA start de CPU (host) een kernel, een functie die door veel GPU-threads tegelijk wordt uitgevoerd. Eén launch maakt een grid van blocks, en elk block bevat veel threads. Elke thread berekent een unieke globale index, meestal i = blockIdx.x × blockDim.x + threadIdx.x, om zijn data-element te kiezen: thread 5 van block 2 met 256 threads per block verwerkt element 517. Omdat threads in hele blocks worden gestart, heeft het laatste block meestal overtollige threads, dus kernels beveiligen geheugentoegangen met if (i < n) om niet voorbij het einde van de array te schrijven.
Elk block wordt toegewezen aan één streaming multiprocessor (SM) en draait daar tot het klaar is, en een SM kan meerdere blocks tegelijk bevatten. Binnen een block worden threads uitgevoerd in warps van 32. De occupancy, de verhouding tussen het aantal aanwezige warps en het maximum van de SM, wordt beperkt door het aantal registers per thread, het shared memory per block en de blockgrootte; bij een lage occupancy zijn er te weinig warps om de geheugenlatentie te verbergen. Blocks draaien onafhankelijk en in willekeurige volgorde, dus binnen een kernel wachten op alle andere blocks kan tot een deadlock leiden; globale synchronisatie wordt normaal verkregen door het werk te verdelen over twee kernellaunches.
Global memory is het DRAM van de GPU: hoge bandbreedte, hoge latentie, en het snelst wanneer naburige threads naburige adressen benaderen. Shared memory is een klein on-chip geheugen per block dat expliciet door het programma wordt beheerd. Het is verdeeld in 32 banks, en threads van een warp die verschillende adressen in dezelfde bank benaderen, worden na elkaar afgehandeld. __syncthreads() laat elke thread van een block wachten tot alle threads zijn aangekomen; als je het in een branch plaatst die sommige threads overslaan, leidt dat tot vastlopers of ongedefinieerde resultaten. Unified memory laat host en device één pointer delen, waarbij pagina's op aanvraag worden gemigreerd.
Kopieën van host naar device gaan over PCI Express en zijn vaak de bottleneck. Streams, geordende wachtrijen van transfers en kernels, laten werk in verschillende streams overlappen, zodat het volgende stuk data kan worden gekopieerd terwijl aan het huidige wordt gerekend. Elke kernellaunch kost ook een paar microseconden, dus programma's die duizenden kleine kernels starten, hebben er baat bij die samen te voegen (fusing), wat ook voorkomt dat tussenresultaten naar het geheugen worden geschreven, of opgenomen launchreeksen in één keer in te dienen met CUDA Graphs. CUDA zelf draait alleen op NVIDIA-GPU's; leveranciersonafhankelijke alternatieven zoals OpenCL volgen een vergelijkbaar model van kernels en work-groups.
Antwoord: B. Een functie die door veel threads tegelijk op de GPU wordt uitgevoerd
Een kernel in GPU-programmering is een functie die vanuit het programma aan de CPU-zijde wordt gestart en waarbij veel threads op de GPU dezelfde code uitvoeren. Elke thread kiest met zijn eigen nummer het data-element waarvoor hij verantwoordelijk is. Bij het optellen van arrays schrijft men de kernel bijvoorbeeld zo dat één thread één element optelt, en start men zoveel threads als er elementen zijn. Het woord is hetzelfde als de kernel van een besturingssysteem, maar de betekenis is anders.
Antwoord: A. Grid → block → thread
Eén keer een kernel starten vormt het gehele grid, het grid is verdeeld in veel blocks en een block in veel threads. De threads van één block worden uitgevoerd op dezelfde SM (een groep rekeneenheden), kunnen via shared memory samenwerken en kunnen synchroniseren. Verschillende blocks worden onafhankelijk uitgevoerd, dus het aantal blocks dat tegelijk draait verandert met de omvang van de GPU, en hetzelfde programma draait op kleine en grote GPU's.
Een block is een door de programmeur gekozen groep threads die op één SM draait en geheugen kan delen en kan synchroniseren. Een warp is de hardwaregroep van 32 threads binnen een block die instructies samen in lockstep uitvoeren.
Gebruik voor een eendimensionale launch i = blockIdx.x * blockDim.x + threadIdx.x. Controleer i < n voordat je arrays benadert, omdat het laatste block vaak meer threads bevat dan er nog elementen over zijn.
Het is snel on-chip geheugen dat wordt gedeeld door de threads van één block, gebruikt om data uit global memory klaar te zetten voor hergebruik en om tussenresultaten uit te wisselen. Anders dan bij een cache bepaalt het programma wat erin komt, en __syncthreads() brengt volgorde aan in de toegangen.
CUDA biedt ook een tool om de occupancy te berekenen.
Welke uitspraak over de bezettingsgraad (occupancy) van een…Ook worden mechanismen gebruikt die een reeks starts samen vastleggen en in één keer versturen (zoals CUDA Graphs).
Een programma dat duizenden keren achter elkaar zeer kleine…CUDA is een mechanisme voor NVIDIA-GPU's, maar OpenCL is een door een branchevereniging vastgestelde standaardspecificatie die kan worden gebruikt op GPU's van meerdere bedrijven, CPU's, FPGA's e.d.
Welk voorbeeld van een programmeermechanisme voor parallell…GPU Shared memory Streaming multiprocessor (SM) Occupancy Via Bank Unified memory Constant memory OpenCL FPGA