Оглавление

• Просто вычисления
• С использованием SDL1.2

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

Это всего лишь справка по тому, как сделать минимальный cuda-файл и скомпилировать его. На этом все.

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