A platform for parallel computing that allows software to access Nvidia GPU for general purpose computing.
The main low level interface lets users to write code in C++ that access parallel core in the GPU.
Referenced from here
Save the following as add.cu.
#include <iostream>
#include <math.h>
// Kernel function to add the elements of two arrays
__global__ void add(int n, float *x, float *y)
{
int index = blockIdx.x * blockDim.x + threadIdx.x;
int stride = blockDim.x * gridDim.x;
for (int i = index; i < n; i += stride)
y[i] = x[i] + y[i];
}
int main(void)
{
int N = 1<<20;
float *x, *y;
// Allocate Unified Memory – accessible from CPU or GPU
cudaMallocManaged(&x, N*sizeof(float));
cudaMallocManaged(&y, N*sizeof(float));
// initialize x and y arrays on the host
for (int i = 0; i < N; i++) {
x[i] = 1.0f;
y[i] = 2.0f;
}
int blockSize = 256;
int numBlocks = (N + blockSize - 1) / blockSize;
// Run kernel on 1M elements on the GPU
add<<<numBlocks, blockSize>>>(N, x, y);
// Wait for GPU to finish before accessing on host
cudaDeviceSynchronize();
// Check for errors (all values should be 3.0f)
float maxError = 0.0f;
for (int i = 0; i < N; i++) {
maxError = fmax(maxError, fabs(y[i]-3.0f));
}
std::cout << "Max error: " << maxError << std::endl;
std::cout << "Value of some element from array: " << y[429801] << std::endl;
// Free memory
cudaFree(x);
cudaFree(y);
return 0;
}
compile and execute using the following:
nvcc add.cu -o add_cuda
./add_cuda
And to profile it to figure out how long each GPU function take to run:
download the nsys_easy script here, then put it in PATH or the current directory, then run
nsys_easy ./add_cuda
or you can use full version:
nsys profile -t cuda --stats=true ./add_cuda
.cu code file with regular C++ codeTo integrate CUDA .cu files into a larger C++ code project, compile the source files seperately and link them
nvcc -c kernel.cu -o kernel.o
g++ -c main.cpp -o main.o
g++ main.o kernel.o -o my_program -L/usr/local/cuda/lib64 -lcudart