Succesorul chipurilor de prelucrare grafică din jocurile aracade, procesoarul grafic, acronim GPU (Graphics Processing Unit), este un circuit electronic, specializat, în crearea și manipularea imaginilor trimise către un afișaj electronic (e.g. monitor).

Utilitatea lui s-a extins ulterior către probleme “embarrassingly parallel”, iar în astăzi sunt folosite la antrenarea antrenarea rețelelor neurale și minarea de criptomonede. Vorbim aici despre întrebuințarea unui GPGPU (General Purpose GPU), un procesor grafic cu o flexibilitate ridicată de programare, capabil de a rezolva și probleme generale.

În execuție, o arhitectură de tip GPU folosește paradigma SIMD (Single Instruction Multiple Data, vezi taxonomia lui Flynn). Acesta presupune:

• schimb rapid de context între thread-uri,
• planificarea în grupuri de thread-uri,
• și orientare către prelucrari masive de date.

Deci, unitățile de tip GPU sunt potrivite pentru paralelismul de date, adică pentru un flux intensiv computațional, cu puține decizii de control.

De obicei, termenul de GPGPU apare atunci când unitatea GPU este folosită ca și coprocesor matematic. Astăzi, majoritatea unităților de tip GPU sunt și GPGPU. Ampla folosire a acestora se datorează:

• diferențelor de putere de procesare brută dintre CPU și GPU (instrucțiuni/secundă)
• standardizarea de API-uri care ușurează munca programatorilor
• răspândirea aplicațiilor ce pot beneficia de pe urma paralelismului de tip SIMD
• cererii pe piața unităților computaționale destinate:
  • atât consumatorilor (PC, Smartphone, TV, etc.),
  • cât și mediilor industriale (Automotive, HPC etc).

Principalii producatori de core-uri IP (intellectual property) tip GPU sunt:

• Intel http://en.wikipedia.org/wiki/List_of_Intel_graphics_processing_units
• Nvidia http://en.wikipedia.org/wiki/List_of_Nvidia_graphics_processing_units
• Amd http://en.wikipedia.org/wiki/List_of_AMD_graphics_processing_units
• Apple https://en.wikipedia.org/wiki/Apple_silicon
• Imagination http://en.wikipedia.org/wiki/List_of_PowerVR_products
• Qualcomm http://en.wikipedia.org/wiki/Adreno
• Vivante http://en.wikipedia.org/wiki/Vivante_Corporation

Dacă un IP de GPU este integrat pe aceeași pastilă de siliciu a unui SoC (System-on-a-Chip), spunem că este un GPU integrat. Exemple de SoC-uri cu IP de GPU integrat includ procesoarele x86 Intel/AMD, cât și majoritatea SoC-urilor pentru dispozitive mobile bazate pe arhitectura ARM (ex. Qualcomm Snapdragon). Un GPU integrat împarte ierarhia de memorie cu alte IP-uri (exp. controllere PCIe/USB/SATA/ETH).

De altfel, un GPU dedicat (discrete GPU) presupune valorificarea unei unui spațiu de memorie, mapat peste VRAM (Video Random-Access Memory), cât și o magistrală PCIe/AGP8x/USB pentru comunicare cu sistemul. Exemple de GPU-uri dedicate sunt seriile de plăci grafice Geforce (Nvidia) și Radeon (AMD).

Exemple de folosire de GPGPU-uri: prelucrări video, audio și de imagini, simulări ale fenomenelor fizice, finanțe, criptografie, design electronic (VLSI), mașini autonome.

• https://www.nvidia.com/en-us/self-driving-cars/partners/bmw/
• https://blogs.nvidia.com/blog/2018/09/18/audi-unveils-e-tron-electric-suv/

Rețele neurale - antrenare vs. inferență.

• https://www.forbes.com/sites/forbestechcouncil/2017/12/01/for-machine-learning-its-all-about-gpus/
• https://www.quora.com/Why-are-GPUs-well-suited-to-deep-learning

Criptomonede - mining via hashing.

• https://coincentral.com/best-gpu-for-mining-2018-edition/

SmartTV, Smartphone - accelerare video, recunoaștere facială/audio.

Simulări fizice - NVIDIA Physx, Folding@Homel

• https://blogs.nvidia.com/blog/2018/11/13/weather-predicted-sc18-gpu-hpc-jensen-huang/

Prelucrări multimedia - filtre imagini GIMP/Photoshop.

Alte domenii - arhivare (WinZip), encriptare.

În cadrul unui sistem ce conține un GPU, procesorul general (CPU) coordonează execuția și este numit “HOST”; pe când unitatea care efectuează calculele (GPU) este numită “DEVICE”.

HOST-ul controlează toate schimbarile de stare în cadrul unui GPU, alocările/transferurile de memorie și evenimentele ce țin de sistemul de operare.

O unitate GPU conține un procesor de comandă (“command processor”) care citește comenzile scrise de către HOST (CPU) în anumite zone de memorie mapate spre access atât către unitatea GPU, cât și către CPU.

În cazul GPU-urilor dedicate, o prelucrare de date necesită în prealabil un transfer din RAM către VRAM. Acest transfer se face printr-o magistrală (PCIe, AGP, USB…). Viteza de transfer RAM-VRAM via magistrală este inferioară vitezei de acces la RAM sau la VRAM. O potențială optimizare în cadrul acestui transfer ar fi intercalarea cu procesarea.

În cazul GPU-urilor integrate transferul RAM↔“VRAM” presupune o mapare de memorie, de multe ori translatată printr-o operație de tip zero-copy.

Programarea unui GPU se face printr-un API (Application Programming Interface). Cele mai cunoscute API-uri orientate către folosirea unui GPU ca coprocesor matematic sunt: CUDA, OpenCL, DirectCompute, OpenACC și Vulkan.

Dezvoltarea de cod pentru laboratoarele de GPU se va face folosind CUDA.

CUDA este un API introdus în 2006 de Nvidia pentru GPU-urile sale. În prezent CUDA este standardul de facto pentru folosirea unităților GPU în industrie și cercetare. Aceasta se datorează faptului că este o platformă stabilă cu multe facilități. Dacă o nouă versiune de CUDA introduce noi funcționalități, dar arhitectura nu le suportă, acestea sunt dezactivate.

În mare toate GPU-urile oferite de Nvidia sunt suportate, diferența fiind la facilitățile suportate. Singura limitare, majoră, a platformei CUDA este că suportă numai unități de procesare de tip GPU de la Nvidia.

Un standard alternativ la CUDA este OpenCL, suportat de Khronos și implementat de majoritatea producătorilor de GPU (inclusiv Nvidia ca o extensie la CUDA). OpenCL suferă de următoarele lipsuri: * suportul este fragmentat * standardul este mult mai restrictiv (decât CUDA) * mai complicat de scris programe (decât CUDA)

CUDA (Compute Unified Device Architecture) permite utilizarea limbajului C pentru programarea pe GPU-urile Nvidia cât și extensii pentru alte limbaje (exp. Python). Deoarece una din zonele țintă pentru CUDA este HPC (High Performance Computing), în care limbajul Fortran este foarte popular, PGI ofera un compilator de Fortran ce permite generarea de cod și pentru GPU-urile Nvidia. Există binding-uri pentru Java (jCuda), Python (PyCUDA) și .NET (CUDA.NET).

Unitatea de bază în cadrul arhitecturii CUDA este numită SM (Streaming Multiprocessor). Ea conține în funcție de generație un număr variabil de CUDA Cores sau SP (Stream Processors) - de regulă între 8SP și 128SP. Unitatea de bază în scheduling este denumită “warp” și este alcatuită din 32 de thread-uri. Vom aborda mai amănunțit arhitectura CUDA în laboratorul următor.

Versiunea de "compute capability" a unui SM, are formatul X.Y, unde X este versiunea majoră, pe când Y este versiunea minoră. Partea majoră identifică generația din care face parte arhitectura.

Partea minoră identifică diferențe incrementale în arhitectură și posibile noi funcționalități.

Știind versiunea majoră și cea minoră cunoaștem facilitățile hardware oferite de către arhitectură.

O listă a GPU-urile NVIDIA și versiunile lor majore/minore se regăsește aici.

CUDA extinde limbajul C prin faptul că permite unui programator să definească funcții C, denumite kernels, care urmează a fi execute în paralel de N thread-uri CUDA. Scopul este de a abstractiza arhitectura GPU astfel încat partea de scheduling cât și gestiunea resurselor se face de catre stiva software CUDA împreună cu suportul hardware. Figura de mai jos denotă distribuirea thread-urilor către două arhitecturi partiționate diferit.

Un kernel se definește folosind specificatorul global iar execuția sa se face printr-o configurație de execuție folosind <<<...>>> . Configurația de execuție denotă numarul de blocuri și numărul de thread-uri dintr-un block. Fiecare thread astfel poate fi identificat unic prin blockIdx și threadIdx.

Mai jos avem definit un kernel, vector_add, care are ca argumente pointeri de tip float, respectiv size_t. Acesta calculează $ f(x) = 2x + 1/(x + 1) $, pentru fiecare element din vector. Numărul total de thread-uri este dimensiunea vectorului.

__global__ void vector_add(const float *a, float *b, const size_t n) {
   // Compute the global element index this thread should process
   unsigned int i = threadIdx.x + blockDim.x * blockIdx.x;
 
   // Avoid accessing out of bounds elements
   if (i < n) {
      b[i] = 2.0 * a[i] + 1.0 / (a[i] + 1.0);
   }
}

Configurația de execuție denotă maparea între date și instrucțiuni. În funcția de kernel, se definește setul de instrucțiuni ce se va executa repetat pe date. Mai jos vector_add este lansat în execuție cu N thread-uri (blocks_no x block_size) organizate câte block_size thread-uri per bloc.

// Launch the kernel
vector_add<<<blocks_no, block_size>>>(device_array_a, device_array_b, num_elements);

#include <stdio.h>
 
__global__ void kernel_example(int value) {
   /**
   * This is a kernel; a kernel is a piece of code that
   * will be executed by each thread from each block in
   * the GPU device.
   */
   printf("[GPU] Hello from the GPU!\n");
   printf("[GPU] The value is %d\n", value);
   printf("[GPU] blockDim=%d, blockId=%d, threadIdx=%d\n",blockDim.x, blockIdx.x, threadIdx.x);
}
 
int main(void) {
   /**
   * Here, we declare and/or initialize different values or we
   * can call different functions (as in every C/C++ program);
   * In our case, here we also initialize the buffers, copy
   * local data to the device buffers, etc (you'll see more about
   * this in the following exercises).
   */
   int nDevices;
   printf("[HOST] Hello from the host!\n");
 
   /**
   * Get the number of compute-capable devices. See more info 
   * about this function in the Cuda Toolkit Documentation.
   */
   cudaGetDeviceCount(&nDevices);
   printf("[HOST] You have %d CUDA-capable GPU(s)\n", nDevices);
 
   /** 
   * Launching the above kernel with a single block, each block
   * with a single thread. The syncrhonize and the checking functions
   * assures that everything works as expected.
   */
   kernel_example<<<1,1>>>(25);
   cudaDeviceSynchronize();
 
   /**
   * Here we can also deallocate the allocated memory for the device
   */
   return 0;
}

O aplicatie CUDA are ca scop executia de cod pe GPU-uri NVIDIA CUDA. In cadrul laboratoarelor partea de CPU (host) va fi folosita exclusiv pentru managementul executiei partii de GPU (device). Aplicatiilor vor viza executia folosind un singur GPU NVIDIA CUDA.

In codul prezentat mai jos, functia vector_add este marcata cu “global” si va fi compilata de catre CUDA NVCC compiler pentru GPU-ul de pe sistem (in cazul cozii xl va fi NVIDIA Pascal P100).

/**
 * This kernel computes the function f(x) = 2x + 1/(x + 1) for each
 * element in the given array.
 */
__global__ void vector_add(const float *a, float *b, const size_t n) {
  // Compute the global element index this thread should process
  unsigned int i = threadIdx.x + blockDim.x * blockIdx.x;
 
  // Avoid accessing out of bounds elements
  if (i < n) {
    b[i] = 2.0 * a[i] + 1.0 / (a[i] + 1.0);
  }
}

Din punct de vedere hardware, partea de host (CPU) are ca memorie principala RAM (chip-uri memorie instalate pe placa de baza via slot-uri memorie) iar partea de device (GPU) are VRAM (chip-uri de memorie prezente pe placa video). Cand vorbim de memoria host (CPU) ne referim la RAM, iar in cazul memoriei device (GPU) la VRAM.

La versiunile mai recente de CUDA, folosind limbajul C/C++, un pointer face referire la spatiul virtual care este unificat pentru host (CPU) si device (GPU). Adresele virtuale insa sunt translatate catre adrese fizice ce rezida ori in memoria RAM (CPU) ori in memoria VRAM (GPU). Astfel este important cum alocam memoria (fie cu malloc pentru CPU sau cudaMalloc pentru GPU) si respectiv sa facem cu atentie transferurile de memorie intre zonele virtuale definite (de la CPU la GPU si respectiv de la GPU la CPU).

// Declare variable to represent ~1M float values and
// computes the amount of bytes necessary to store them
const int num_elements = 1 << 16;
const int num_bytes = num_elements * sizeof(float);
 
// Declaring the 'host arrays': a host array is the classical
// array (static or dynamically allocated) we worked before.
float *host_array_a = 0;
float *host_array_b = 0;
 
// Declaring the 'device array': this array is the equivalent
// of classical array from C, but specially designed for the GPU
// devices; we declare it in the same manner, but the allocation
// process is going to be different
float *device_array_a = 0;
float *device_array_b = 0;

Functia malloc va intoarce o adresa virtuala ce va avea corespondent o adresa fizica din RAM.

// Allocating the host array
host_array_a = (float *) malloc(num_bytes);
host_array_b = (float *) malloc(num_bytes);

Functia cudaMalloc va intoarce o adresa virtuala ce va avea corespondent o adresa fizica din VRAM.

// Allocating the device's array; notice that we use a special
// function named cudaMalloc that takes the reference of the
// pointer declared above and the number of bytes.
cudaMalloc((void **) &device_array_a, num_bytes);
cudaMalloc((void **) &device_array_b, num_bytes);
 
// If any memory allocation failed, report an error message
if (host_array_a == 0 || host_array_b == 0|| device_array_a == 0 || device_array_b == 0) {
  printf("[HOST] Couldn't allocate memory\n");
  return 1;
}

// Initialize the host array by populating it with float values  
for (int i = 0; i < num_elements; ++i) {
  host_array_a[i] = (float) i;
}
 
// Copying the host array to the device memory space; notice the
// parameters of the cudaMemcpy function; the function default
// signature is cudaMemcpy(dest, src, bytes, flag) where
// the flag specifies the transfer type.
//
// host -> device: cudaMemcpyHostToDevice
// device -> host: cudaMemcpyDeviceToHost
// device -> device: cudaMemcpyDeviceToDevice
cudaMemcpy(device_array_a, host_array_a, num_bytes, cudaMemcpyHostToDevice);

// Compute the parameters necessary to run the kernel: the number
// of blocks and the number of threads per block; also, deal with
// a possible partial final block
const size_t block_size = 256;
size_t blocks_no = num_elements / block_size;
 
if (num_elements % block_size) 
  ++blocks_no;
 
// Launch the kernel
vector_add<<<blocks_no, block_size>>>(device_array_a, device_array_b, num_elements);
cudaDeviceSynchronize();

// Copy the result back to the host memory space
cudaMemcpy(host_array_b, device_array_b, num_bytes, cudaMemcpyDeviceToHost);
 
// Print out the first 10 results
for (int i = 0; i < 10; ++i) {
  printf("Result %d: 2 * %1.1f + 1.0/(%1.1f + 1.0)= %1.3f\n", 
  i, host_array_a[i], host_array_a[i], host_array_b[i]);
}

// Deallocate memory
free(host_array_a);
free(host_array_b);
cudaFree(device_array_a);
cudaFree(device_array_b);

Desi pentru un programator partile de host/CPU respectiv device/GPU pot fi in acelasi fisier *.cu, compilatorul CUDA (nvcc) le separa facand o compilare diferita pentru partea de host/CPU respectiv device/GPU. Figura de mai jos denota acest aspect.

Intrati pe frontend-ul fep8.grid.pub.ro folosind contul de pe curs.upb.ro. Executati comanda

srun --pty -p xl --gres gpu:1 /bin/bash

pentru a accesa una din statiile cu GPU-uri. Cozile ce au unitati GPU NVIDIA sunt xl si ucsx.

[@fep8 ~]$ srun --pty -p xl --gres gpu:1 /bin/bash
[@wn-xyz ~]$ nvidia-smi       # NVIDIA System Management Interface program
+---------------------------------------------------------------------------------------+
| NVIDIA-SMI 535.104.05             Driver Version: 535.104.05   CUDA Version: 12.2     |
|-----------------------------------------+----------------------+----------------------+
| GPU  Name                 Persistence-M | Bus-Id        Disp.A | Volatile Uncorr. ECC |
| Fan  Temp   Perf          Pwr:Usage/Cap |         Memory-Usage | GPU-Util  Compute M. |
|                                         |                      |               MIG M. |
|=========================================+======================+======================|
|   0  Tesla P100-PCIE-16GB           On  | 00000000:0D:00.0 Off |                    0 |
| N/A   28C    P0              27W / 250W |      0MiB / 16384MiB |      0%      Default |
|                                         |                      |                  N/A |
+-----------------------------------------+----------------------+----------------------+
 
+---------------------------------------------------------------------------------------+
| Processes:                                                                            |
|  GPU   GI   CI        PID   Type   Process name                            GPU Memory |
|        ID   ID                                                             Usage      |
|=======================================================================================|
|  No running processes found                                                           |
+---------------------------------------------------------------------------------------+

Pentru laboratoarele de GP-GPU Computing vom folosi CUDA 10.2 sau 11.4 aici.

SDK-ul CUDA de la NVidia include atat implementarea de CUDA API cat si cea de OpenCL API. In cadrul laboratoarelor vom programa numai folosind CUDA. Verificam mai jos ca scheletul laboratorului compileaza.

[@wn-xyz ~]$ wget -O lab7_skl.tar.gz https://ocw.cs.pub.ro/courses/_media/asc/lab7/lab7_skl.tar.gz
--2020-03-22 18:52:14--  http://ocw.cs.pub.ro/courses/_media/asc/lab7/lab7_skl.tar.gz
Resolving ocw.cs.pub.ro (ocw.cs.pub.ro)... 141.85.227.65
Connecting to ocw.cs.pub.ro (ocw.cs.pub.ro)|141.85.227.65|:80... connected.
HTTP request sent, awaiting response... 302 Found
Location: https://ocw.cs.pub.ro/courses/_media/asc/lab7/lab7_skl.tar.gz [following]
--2020-03-22 18:52:14--  https://ocw.cs.pub.ro/courses/_media/asc/lab7/lab7_skl.tar.gz
Connecting to ocw.cs.pub.ro (ocw.cs.pub.ro)|141.85.227.65|:443... connected.
HTTP request sent, awaiting response... 200 OK
Length: 4884 (4.8K) [application/octet-stream]
Saving to: 'lab7_skl.tar.gz'
 
100%[==================================================================================================>] 4,884       --.-K/s   in 0s      
 
2020-03-22 18:52:14 (11.1 MB/s) - 'lab7_skl.tar.gz' saved [4884/4884]
 
[@wn-xyz ~]$ tar -xvzf lab7_skl.tar.gz
task1/
task1/task1.cu
task1/Makefile
task1/Makefile_Cluster
task2/
task2/task2.cu
task2/Makefile
task2/Makefile_Cluster
...
[@wn-xyz ~]$ cd task1/

[@wn-xyz ~]$ apptainer run --nv docker://gitlab.cs.pub.ro:5050/asc/asc-public/cuda-labs:1.9.1
INFO:    Using cached SIF image
Apptainer>

În cadrul imaginei CUDA avem compilatorul nvcc.

COMPILER=nvcc
LIBS=-lm
 
%: %.cu
	$(COMPILER) $^ -o $@ $(LIBS)
 
clean:
	rm -rf task0

Exemplu de compilare si rulare interactiva pe coada hpsl

Apptainer> make task1
nvcc task1.cu -o task1 -lm
Apptainer> ./task1
[HOST] Hello from the host!
[HOST] You have 3 CUDA-capable GPU(s)
Apptainer> make clean
rm -rf task1
Apptainer> cd ../task2
Apptainer> make task2
nvcc task2.cu -o task2 -lm
Apptainer> ./task2
Result 0: 2 * 0.0 + 1.0/(0.0 + 1.0)= 0.000
Result 1: 2 * 1.0 + 1.0/(1.0 + 1.0)= 0.000
Result 2: 2 * 2.0 + 1.0/(2.0 + 1.0)= 0.000
Result 3: 2 * 3.0 + 1.0/(3.0 + 1.0)= 0.000
Result 4: 2 * 4.0 + 1.0/(4.0 + 1.0)= 0.000
Result 5: 2 * 5.0 + 1.0/(5.0 + 1.0)= 0.000
Result 6: 2 * 6.0 + 1.0/(6.0 + 1.0)= 0.000
Result 7: 2 * 7.0 + 1.0/(7.0 + 1.0)= 0.000
Result 8: 2 * 8.0 + 1.0/(8.0 + 1.0)= 0.000
Result 9: 2 * 9.0 + 1.0/(9.0 + 1.0)= 0.000
Apptainer> 

Exemplu executie program CUDA folosind rularea ne-interactiva pe coada de executia hpsl (apelul trebuie facut de pe fep8.grid.pub.ro):

[fep8 ~]$ cd task2/
[fep8 task2]$ cat ../utils/batch_run.sh 
#!/bin/bash
apptainer exec --nv $CONTAINER_IMAGE  \
./$TASK
[fep8 task2]$ sbatch --time 01:00:00 -p xl --gres gpu:1 --export=TASK=task2,CONTAINER_IMAGE=docker://gitlab.cs.pub.ro:5050/asc/asc-public/cuda-labs:1.9.1 ../utils/batch_run.sh
Submitted batch job 1816
[fep8 task2]$ squeue | grep 1816
              1816      hpsl batch_ru stefan_d  R       0:07      1 hpsl-wn01
[fep8 task2]$ cat slurm-1816.out 
INFO:    Using cached SIF image
Result 0: 2 * 0.0 + 1.0/(0.0 + 1.0)= 1.000
Result 1: 2 * 1.0 + 1.0/(1.0 + 1.0)= 2.500
Result 2: 2 * 2.0 + 1.0/(2.0 + 1.0)= 4.333
Result 3: 2 * 3.0 + 1.0/(3.0 + 1.0)= 6.250
Result 4: 2 * 4.0 + 1.0/(4.0 + 1.0)= 8.200
Result 5: 2 * 5.0 + 1.0/(5.0 + 1.0)= 10.167
Result 6: 2 * 6.0 + 1.0/(6.0 + 1.0)= 12.143
Result 7: 2 * 7.0 + 1.0/(7.0 + 1.0)= 14.125
Result 8: 2 * 8.0 + 1.0/(8.0 + 1.0)= 16.111
Result 9: 2 * 9.0 + 1.0/(9.0 + 1.0)= 18.100

Puteți folosi Makefile_Cluster:

[fep8 ~]$ cd task2/
[fep8 task2]$ make -f Makefile_Cluster clean
rm -rf task2
rm -rf slurm-*
[fep8 task2]$ make -f Makefile_Cluster task2
sbatch --time 01:00:00 -p hpsl --gres gpu:1 --export=TASK=task2,CONTAINER_IMAGE=docker://gitlab.cs.pub.ro:5050/asc/asc-public/cuda-labs:1.9.1 ../utils/batch_build.sh | ../utils/batch_wait.sh
INFO:    Using cached SIF image
nvcc task2.cu -o task2 -lm
[fep8 task2]$ make -f Makefile_Cluster run_task1
sbatch --time 01:00:00 -p xl --export=TASK=task2,CONTAINER_IMAGE=docker://gitlab.cs.pub.ro:5050/asc/asc-public/cuda-labs:1.9.1 ../utils/batch_run.sh | ../utils/batch_wait.sh
INFO:    Using cached SIF image
Result 0: 2 * 0.0 + 1.0/(0.0 + 1.0)= 1.000
Result 1: 2 * 1.0 + 1.0/(1.0 + 1.0)= 2.500
Result 2: 2 * 2.0 + 1.0/(2.0 + 1.0)= 4.333
Result 3: 2 * 3.0 + 1.0/(3.0 + 1.0)= 6.250
Result 4: 2 * 4.0 + 1.0/(4.0 + 1.0)= 8.200
Result 5: 2 * 5.0 + 1.0/(5.0 + 1.0)= 10.167
Result 6: 2 * 6.0 + 1.0/(6.0 + 1.0)= 12.143
Result 7: 2 * 7.0 + 1.0/(7.0 + 1.0)= 14.125
Result 8: 2 * 8.0 + 1.0/(8.0 + 1.0)= 16.111
Result 9: 2 * 9.0 + 1.0/(9.0 + 1.0)= 18.100

mkdir /mnt/asc
sudo chown -R <user> /mnt/asc
decomentam linia 'user_allow_other' din /etc/fuse.conf
sshfs -o allow_other <user.moodle>@fep8.grid.pub.ro:/ /mnt/asc

Pentru inceput:

1. Logati-va pe fep8.grid.pub.ro folosind contul de pe curs.upb.ro
2. Executați comanda:

wget -O lab7_skl.tar.gz http://ocw.cs.pub.ro/courses/_media/asc/lab7/lab7_skl.tar.gz

1. Dezarhivati folosind comanda tar -xzvf lab7_skl.tar.gz

Modificarile se vor face acolo unde este necesar in task_<i>.cu unde <i> este numarul taskului. Urmariti indicatiile TODO din cod. De asemenea, va recomandam sa folositi documentatia oficiala CUDA Toolkit Documentation de la adresa: https://docs.nvidia.com/cuda/. Aici veti gasi informatii despre majoritatea functiilor de care aveti nevoie (folositi functia search).

Task 1 - Rulați task1 ca exemplu pentru a verifica funcționalitatea CUDA pe GPU

Task 2 - Rulați task2 ca exemplu pentru efectuarea unor operații pe GPU

Task 4 - Efectuați adunarea a doi vectori folosind CUDA în task4.cu

• Sugestia este de a face intai taskul 4 si apoi taskul 3 pentru ca are sens dpdv logic - e mai usor de inteles ce se intampla.

Task 3 - Urmăriți TODO–uri din taks3.cu

• Listați informații despre device-urile existente și selectați primul device
• Completați și rulați kernelul kernel_parity_id
• Completați și rulați kernelul kernel_block_id; explicați rezultatul
• Completați și rulați kernelul kernel_thread_id; explicați rezultatul

Task 5 - Urmăriți instrucțiunile din task5.cu pentru a realiza interschimbarea a doi vectori

Schelet Laborator 7

Enunt Laborator 7

• Responsabili laborator: Grigore Lupescu, Ștefan-Dan Ciocîrlan, Costin Carabaș

• Documentatie CUDA:
  • CUDA C Programming
  • CUDA NVCC compiler
  • CUDA Visual Profiler
  • CUDA 9.1 Toolkit
  • CUDA GPUs
• Acceleratoare hpsl (hpsl-wn01, hpsl-wn02, hpsl-wn03)
  • NVIDIA Tesla K40M
  • NVIDIA Tesla
• Acceleratoare dp (dp-wn01, dp-wn02, dp-wn03, dp-wn04)
  • NVIDIA Tesla C2070
  • NVIDIA Tesla 2050/2070
  • NVIDIA CUDA Fermi/Tesla
• Advanced CUDA
  • CUDA Streams
  • CUDA Dynamic Parallelism

• Extra
  • CUDA Thread Basics
  • An Even Easier Introduction to CUDA
  • Supercomputing 2011 Tutorial