Лисья Нора
09:14
Главная » C/C++ » CUDA NVCC

Оглавление


§ Просто вычисления

Это всего лишь справка по тому, как сделать минимальный cuda-файл и скомпилировать его. На этом все.
Но, перед тем как запускать, надо обязательно поставить NVidia CUDA Toolkit.
nvcc cuda.cu -o cuda && ./cuda
И теперь сам код:
#include <cuda_runtime_api.h>
#include <stdio.h>
 
void get_device_info()
{
int deviceCount;
cudaDeviceProp deviceProp;
 
//Сколько устройств CUDA установлено на PC.
cudaGetDeviceCount(&deviceCount);
 
printf("Device count: %d\n\n", deviceCount);
 
for (int i = 0; i < deviceCount; i++)
{
//Получаем информацию об устройстве
cudaGetDeviceProperties(&deviceProp, i);
 
//Выводим иформацию об устройстве
printf("Device name: %s\n", deviceProp.name);
printf("Total global memory: %d\n", (int)deviceProp.totalGlobalMem);
printf("Shared memory per block: %d\n", (int)deviceProp.sharedMemPerBlock);
printf("Registers per block: %d\n", deviceProp.regsPerBlock);
printf("Warp size: %d\n", deviceProp.warpSize);
printf("Memory pitch: %d\n", (int)deviceProp.memPitch);
printf("Max threads per block: %d\n", deviceProp.maxThreadsPerBlock);
 
printf("Max threads dimensions: x = %d, y = %d, z = %d\n",
deviceProp.maxThreadsDim[0],
deviceProp.maxThreadsDim[1],
deviceProp.maxThreadsDim[2]);
 
printf("Max grid size: x = %d, y = %d, z = %d\n",
deviceProp.maxGridSize[0],
deviceProp.maxGridSize[1],
deviceProp.maxGridSize[2]);
 
printf("Clock rate: %d\n", deviceProp.clockRate);
printf("Total constant memory: %d\n", (int)deviceProp.totalConstMem);
printf("Compute capability: %d.%d\n", deviceProp.major, deviceProp.minor);
printf("Texture alignment: %d\n", (int)deviceProp.textureAlignment);
printf("Device overlap: %d\n", deviceProp.deviceOverlap);
printf("Multiprocessor count: %d\n", deviceProp.multiProcessorCount);
 
printf("Kernel execution timeout enabled: %s\n",
deviceProp.kernelExecTimeoutEnabled ? "true" : "false");
}
}
 
__global__ void SomeKernel(int* data, int length)
{
unsigned int threadId = blockIdx.x * blockDim.x + threadIdx.x;
 
if (threadId < length)
{
data[threadId] = 1;
}
}

§ Главный файл

int main()
{
get_device_info();
 
int w = 16;
int h = 16;
int length = w * h;
 
// Выделение оперативной памяти (для CPU)
int* hostData = (int*)malloc(length * sizeof(int));
 
// Инициализация исходных данных
for (int i = 0; i < length; i += 2)
{
hostData[i] = 2;
hostData[i+1] = 1;
}
 
// Выделение памяти GPU
int* deviceData;
 
cudaMalloc((void**)&deviceData, length * sizeof(int));
 
// Копирование исходных данных в GPU для обработки
cudaMemcpy(deviceData, hostData, length * sizeof(int), cudaMemcpyHostToDevice);
 
// http://choorucode.com/2011/02/16/cuda-dim3/
dim3 threads = dim3(w);
dim3 blocks = dim3(h);
 
// Запуск ядра из (length / 256) блоков по 256 потоков,
// предполагая, что length кратно 256
SomeKernel<<<blocks, threads>>>(deviceData, length);
 
// Считывание результата из GPU
cudaMemcpy(hostData, deviceData, length * sizeof(int), cudaMemcpyDeviceToHost);
 
// Отображение результата
for (int i = 0; i < length; i++)
{
if (i % 16 == 0) printf("\n");
printf("%d\t", (int)hostData[i]);
}
 
printf("\n");
 
return 0;
}

§ С использованием SDL1.2

В этом коде создается SDL-окно и ничего не происходит с ним, просто показывается, что можно так сделать.
#include <stdio.h>
#include <SDL.h>
 
__global__
void saxpy(int n, float a, float *x, float *y)
{
int i = blockIdx.x * blockDim.x + threadIdx.x;
if (i < n) y[i] = cos(sin(a*x[i]) + y[i]);
}
 
int main(void)
{
// 256 млн
int N = 1 << 27;
 
SDL_Init(SDL_INIT_VIDEO | SDL_INIT_TIMER);
SDL_EnableUNICODE(1);
SDL_Surface* screen = SDL_SetVideoMode(320, 200, 32, SDL_HWSURFACE | SDL_DOUBLEBUF);
SDL_WM_SetCaption("CUDA NVCC", 0);
 
float *x, *y, *d_x, *d_y;
 
// Выделить память на CPU
x = (float*) malloc(N*sizeof(float));
y = (float*) malloc(N*sizeof(float));
 
// Выделить память на GPU
cudaMalloc(&d_x, N*sizeof(float));
cudaMalloc(&d_y, N*sizeof(float));
 
Uint32 t1 = SDL_GetTicks();
printf("START\n");
 
for (int i = 0; i < N; i++) {
x[i] = 1.0f;
y[i] = 2.0f;
}
 
Uint32 t2 = SDL_GetTicks();
printf("FILL :: %d\n", t2 - t1);
 
// Скопировать данные
cudaMemcpy(d_x, x, N*sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(d_y, y, N*sizeof(float), cudaMemcpyHostToDevice);
 
Uint32 t3 = SDL_GetTicks();
printf("COPY x 2 :: %d\n", t3 - t2);
 
// Perform SAXPY on elements
saxpy<<<(N+255) / 256, 256>>>(N, 2.0f, d_x, d_y);
 
// Подождать пока завершится
cudaDeviceSynchronize();
 
Uint32 t4 = SDL_GetTicks();
printf("COMPUTE :: %d\n", t4 - t3);
 
// Скопировать обратно
cudaMemcpy(y, d_y, N*sizeof(float), cudaMemcpyDeviceToHost);
 
Uint32 t5 = SDL_GetTicks();
printf("COPY BACK :: %d\n", t5 - t4);
 
cudaFree(d_x);
cudaFree(d_y);
free(x);
free(y);
}
Код для компиляции
nvcc main.cu `sdl-config --cflags --libs` -lSDL -lm -o main