Pertanyaan pertama yang muncul adalah, apa itu CUDA? Dan apa hubungannya dengan Nvidia? CUDA adalah platform parallel computing dan programming oleh Nvidia. parallel computing? ya, disini kita akan memperdayakan core dari GPU yang sangat banyak jika dibandingkan dengan core dari CPU (dalam konteks ini Nvidia) dalam melakukan komputasi yang kita inginkan.
 |
| perbandingan core CPU dan GPU |
Programming? Ya, untuk "memerintah" kita butuh "surat perintah" (baca: source code) yang akan di rancang terlebih dahulu. Bahasa program yang digunakan untuk CUDA adalah C/C++ dan Fortran tapi kali ini akan fokus ke C. Bahasa C yang digunakan dalam programming cuda bukan C biasa, tapi versi ekstensi dari CUDA untuk keperluan memerintah GPU.
 |
| perbedaan bahasa C biasa dan versi ekstensinya dengan CUDA |
Cukup dengan perkenalannya, kita beralih ke persiapan untuk programming CUDA. Saya pakai laptop ASUS A456U dengan OS Linux Ubuntu 16.04 dalam praktek kali ini. Pertama yang harus disiapkan adalah drivernya, untuk itu ikuti petunjuk pada link ini. Setelah itu, install CUDA dengan
$ sudo dpkg -i cuda-repo-ubuntu1604_7.5-18_amd64.deb
$ sudo apt-get update
$ sudo apt install nvidia-cuda-toolkit
Setelah itu, jika anda mungkin nyaman dengan IDE bisa menginstall IDE dari CUDA yaitu
$ sudo apt install nvidia-nsight
Tapi saya sendiri lebih nyaman dengan text editor andalan saya yaitu Geany, mungkin sudah kebiasaaan.
Ok, ayo mulai dengan koding hello world, yang terlihat seperti
__global__ void mykernel(void) {
}
int main(void) {
mykernel<<<1,1>>>();
printf("Hello World!\n");
return 0;
}
Save dengan nama hellocuda.cu dan untuk mencompile dan menjalankannya cukup dengan
$ nvcc hellocuda.cu -o hellocuda
./hellocuda
Seperti yang terlihat di source code hellocuda.cu kita ada beberapa baris kode yang asing yaitu dibaris pertama ada __global__ dan di baris 4 ada mykernel<<<1,1>>>();.
Jadi sebelumnya, pada source code di atas yang sudah merupakan ektensi dari CUDA, baris kode dibedakan menjadi 2, yaitu device code dan host code. Device code adalah baris kode yang akan di eksekusi oleh GPU, yang di tandai dengan awal fungsinya ada __global__, sedangkan host code adalah baris kode biasa. Jadi, device code dan host code disini selalu juga bisa disebut host function atau device function karena selalu merupakan satu function yang utuh, entah itu host code maupun device code.
Kemudian ada mykernel<<<1,1>>>();. Ini adalah baris kode untuk mengeksekusi device code. 1,1 pada menandakan jumlah block dan thread yang diinginkan untuk eksekusi device code mykernel, syntaxnya seperti
mykernel<<<jumlah_block,jumlah_thread>>>();
Dimana jumlah_block dan jumlah_thread disini masing-masing mewakilkan jumlah block dan jumlah thread yang dinginkan untuk eksekusi device code. Block? Thread? Untuk menggambarkan block dan thread mungkin seperti gambar dibawah ini.
 |
| Terlihat disini kumpulan block disebut grids dan dalam 1 block terdapat beberapa thread. |
Selain itu, thread pada block yang sama bisa saling share memory yang mana thread pada block berbeda tidak akan aware pada itu. Untuk lebih paham berikut source codenya.
#include <stdio.h>
#define N 512
#define TRT 16
__global__ void mykernel(int *a, int *b, int *c) {
int index;
index = threadIdx.x + blockIdx.x * blockDim.x;
c[index] = a[index] + b [index];
}
void random_ints(int* a, int x)
{
int i;
for (i = 0; i < x; ++i)
a[i] = rand();
}
int main(void) {
int *a, *b, *c;
int *d_a, *d_b, *d_c;
int size = N * sizeof(int);
cudaMalloc((void **)&d_a, size);
cudaMalloc((void **)&d_b, size);
cudaMalloc((void **)&d_c, size);
a = (int *) malloc(size); random_ints(a, N);
b = (int *) malloc(size); random_ints(b, N);
c = (int *) malloc(size);
// Copy inputs to device
cudaMemcpy(d_a, a, size, cudaMemcpyHostToDevice);
cudaMemcpy(d_b, b, size, cudaMemcpyHostToDevice);
mykernel<<<N / TRT, TRT>>>(d_a, d_b, d_c);
// Copy result back to host
cudaMemcpy(c, d_c, size, cudaMemcpyDeviceToHost);
for (int i = 0; i < N; i++)
printf("result ke %d : %d\n", i, c[i]);
// Cleanup
free(a); free(b); free(c);
cudaFree(d_a); cudaFree(d_b); cudaFree(d_c);
return 0;
}
Wah, sepertinya banyak harus di jelaskan dari kode di atas karena banyak function asing seperti cudaMalloc, cudaMemcpy, cudaFree dan flag serta lain-lainnya. Ok, ada tambahan pada function mykernel kita yaitu
int index;
index = threadIdx.x + blockIdx.x * blockDim.x;
c[index] = a[index] + b [index];
Dimana threadIdx.x, blockIdx.x dan blockDim.x disini masing-masing artinya adalah thread index, block index dan dimension block (banyaknya thread dalam satu block).
Pada kode ini, masing-masing thread pada masing-masing block akan mengurus operasi penjumlahan (c = a + b) dari tiap-tiap nomor indeks pada array secara parallel. Untuk itulah terlebih dahulu jumlah thread yang akan beroperasi harus di atur sesuai dengan ukuran array pada kode di atas.
Diketahui ukuran array pada kode di atas adalah 512 yang berarti kita butuh 512 thread. Tiap block akan mempunyai jumlah thread yang sama artinya salah satu kombinasi block dan thread yang memungkinkan untuk kode kali ini adalah
512 = 16 x 32
dimana jumlah block 32 dan pada masing-masing block akan punya 15 thread.
mykernel<<<N / TRT, TRT>>>(d_a, d_b, d_c);
Sebenarnya kita bisa saja langsung deklarasikan 1 block dan 512 thread
mykernel<<<1, N>>>(d_a, d_b, d_c);
Atau 512 block 1 thread
mykernel<<<N, 1>>>(d_a, d_b, d_c);
Tapi untuk mendapat pemahaman mengenai hubungan thread dan block maka di buat seperti demikian.
index pada function mykernel membantu program mengetahui bagian array mana yang harus di kerjakan suatu thread tersebut dengan mencari tahu terlebih dahulu index thread dan blocknya masing-masing.
int index;
index = threadIdx.x + blockIdx.x * blockDim.x;
c[index] = a[index] + b [index];
cudaMalloc, cudaMemcpy, cudaFree ini analog dengan malloc, free, dan memcpy tapi mereka berurusan dengan device memory (memori yang dikelola GPU).
Pada cudaMemcpy ada flag asing seperti cudaMemcpyHostToDevice dan cudaMemcpyDeviceToHost. Flag-flag itu sebenarnya self explanatory karena sesuai namanya pada cudaMemcpyHostToDevice maksudnya adalah untuk mengarahkan cudaMemcpy untuk melakukan copy dari memori host ke memori device dan begitu juga dengan cudaMemcpyDeviceToHost yaitu untuk melakukan copy dari memori device ke host. Agar bisa di tampilkan outputnya, hasil perhitungan harus selalu di konversi ke memori host.
cudaMemcpy(c, d_c, size, cudaMemcpyDeviceToHost);
for (int i = 0; i < N; i++)
printf("result ke %d : %d\n", i, c[i]);
Mungkin cukup sekian untuk permulaan singkatnya. Masih ada beberapa hal yang belum dikenalkan seperti shared memory yang merupakan keuntungan dari penggunaan thread yang banyak pada masing-masing block atau sistem block index yang seperti matriks, tapi ini saja saya rasa cukup untuk perkenalan awal karena semakin banyak materi yang coba dipahami sekaligus maka akan semakin banyak pula yang tidak dipahami, slow but sure. Dilain kesempatan mungkin akan dibahas.
referensi:
http://ubuntuhandbook.org/index.php/2016/04/switch-intel-nvidia-graphics-ubuntu-16-04/
https://devblogs.nvidia.com/parallelforall/using-shared-memory-cuda-cc/
https://developer.nvidia.com/cuda-zone
http://www.nvidia.com/object/what-is-gpu-computing.html
https://blogs.nvidia.com/blog/2012/09/10/what-is-cuda-2/
http://docs.nvidia.com/cuda/cuda-getting-started-guide-for-linux/#axzz4JDp4rJPx
http://15418.courses.cs.cmu.edu/spring2013/article/11
Komentar
Posting Komentar