nano main.cu<\/code><\/pre>\n\n\n\n#include <stdio.h>\n#include <stdlib.h><\/code><\/pre>\n\n\n\n2. Get GPU card information<\/h3>\n\n\n\nint main()\n{\n cudaError_t cuErr;\n int cuDev;\n struct cudaDeviceProp cuDevProp;\n cuErr = cudaGetDevice(&cuDev);\n if(cuErr == cudaSuccess){\n printf(\"Device %d\\r\\n\",cuDev);\n cuErr = cudaGetDeviceProperties(&cuDevProp,cuDev);\n if(cuErr == cudaSuccess){\n printf(\"Device Properties:\\r\\n\");\n printf(\"- Name : %s\\r\\n\",cuDevProp.name);\n printf(\"- Total Global Mem : %lu\\r\\n\",cuDevProp.totalGlobalMem);\n printf(\"- Shared Mem Per Block : %lu\\r\\n\",cuDevProp.sharedMemPerBlock);\n printf(\"- Max Threads Per Block: %d\\r\\n\",cuDevProp.maxThreadsPerBlock);\n printf(\"- Compute Mode : %d\\r\\n\",cuDevProp.computeMode);\n }\n else{\n printf(\"Failed to get Device Properties!\\r\\n\");\n }\n }\n else{\n printf(\"Failed to get Device!\\r\\n\");\n }\n return 0;\n}<\/code><\/pre>\n\n\n\nCompile and test<\/strong><\/p>\n\n\n\nnvcc main.cu\n.\/a.out<\/code><\/pre>\n\n\n\nResult<\/strong><\/p>\n\n\n\nltkhanh@ServerTX:~\/cuda$ .\/a.out \nDevice 0\nDevice Properties:\n- Name : GeForce GT 630\n- Total Global Mem : 2081619968\n- Shared Mem Per Block : 49152\n- Max Threads Per Block: 1024\n- Compute Mode : 0 <\/code><\/pre>\n\n\n\n3. Terminology<\/h3>\n\n\n\n- Host<\/strong>: the CPU and its memory (host memory)<\/li>
- Device<\/strong>: the GPU and its memory (device memory)<\/li><\/ul>\n\n\n\n
read more at https:\/\/www.nvidia.com\/docs\/IO\/116711\/sc11-cuda-c-basics.pdf<\/a><\/p>\n\n\n\n4. Device code<\/h3>\n\n\n\n__global__ void mykernel(void) {\n printf(\"[Device] Hello Host from Cuda block %d, thread %d\\n\", blockIdx.x, threadIdx.x);\n}<\/kbd><\/code><\/pre>\n\n\n\n5. Host code<\/h3>\n\n\n\nint main(void) {\n mykernel<<<1,1>>>();\n cudaDeviceSynchronize();\n printf(\"[Host] Device error: %s\\n\", cudaGetErrorString(cudaGetLastError()));\n return 0;\n}<\/code><\/pre>\n\n\n\nLine 2 calls device function mykernel.<\/em> ‘<<<‘ and ‘>>>’ tell that we are calling from host to device.
Generic format is <function name><<<N<\/strong>,K<\/strong>>>><\/em> with N is the number of blocks in parallel, K is the number of threads per block.<\/p>\n\n\n\nLine 3 blocks until the device has completed all preceding requested tasks, that allows the printf in the device function printing results to the host terminal. Without this line, we will see no output.<\/p>\n\n\n\n
In line 4, we get the last error (if have) while calling Cuda functions and convert it to string to print to terminal.<\/p>\n\n\n\n
6. Compile and test<\/h3>\n\n\n\nnvcc -arch sm_21 -o test main.cu<\/code><\/pre>\n\n\n\nOption -arch tells NVCC which GPU architecture we are using (sm_21 or Fermi architecture in this case). Option -o to set the output filename. Now we execute the binary file<\/p>\n\n\n\n
ltkhanh@ServerTX:~\/cuda$ .\/test \n[Device] Hello Host from Cuda block 0, thread 0\n[Host] Device error: no error<\/code><\/pre>\n\n\n\n7. Transfer data between host and device<\/h3>\n\n\n\n
Host and device memory are separate entities. To transfer data between host and device, we need pointers for each memory:<\/p>\n\n\n\n
- Device<\/em><\/strong> pointers point to GPU memory. They may be passed to\/from host code and may not be dereferenced in host code.<\/li>
- Host<\/em><\/strong> pointers point to CPU memory. They may be passed to\/from device code and may not be dereferenced in device code.<\/li><\/ul>\n\n\n\n
![\"\"](\"https:\/\/bigdolphin.com.vn\/wp-content\/uploads\/2022\/05\/image.png\")
<\/figure><\/div>\n\n\n\n