Accelerated Ray Tracing （一）

https://developer.nvidia.com/blog/easy-introduction-cuda-c-and-c/

host 指的是CPU及其内存，device指的是GPU及其内存。

在主机上运行的代码可以管理host 和device上的内存，还可以启动kernels ，这些kernels 是在device上执行的functions。这些kernels 由许多GPU线程并行执行。

• 声明和分配主host 和device内存。
• 初始化host 数据。
• 将数据从host 传输到device。
• 执行一个或多个kernels 。
• 将结果从device传输到host 。

__global__ ：声明device kernel functions;

device variables blockDim, blockIdx, and threadIdx 获取全局索引

#include <stdio.h>

__global__
void saxpy(int n, float a, float* x, float* y)//kernel that runs in parallel on the GPU,
{
    //Device Code
    int i = blockIdx.x * blockDim.x + threadIdx.x;//generates a global index 
    if (i < n) y[i] = a * x[i] + y[i];//n, to ensure there are no out-of-bounds memory accesses.
}

int main(void)//host code
{
    //声明两组数组
    int N = 1 << 20;
    float* x, * y, * d_x, * d_y;
    x = (float*)malloc(N * sizeof(float));//  The pointers x and y point to the host arrays, allocated with malloc in the typical fashion,
    y = (float*)malloc(N * sizeof(float));

    cudaMalloc(&d_x, N * sizeof(float));//d_x and d_y arrays point to device arrays allocated with the cudaMalloc function from the CUDA runtime API.
    cudaMalloc(&d_y, N * sizeof(float));

    for (int i = 0; i < N; i++) {//initializes the host arrays
        x[i] = 1.0f;
        y[i] = 2.0f;
    }

    cudaMemcpy(d_x, x, N * sizeof(float), cudaMemcpyHostToDevice);//initialize the device arrays
    cudaMemcpy(d_y, y, N * sizeof(float), cudaMemcpyHostToDevice);

    // Perform SAXPY on 1M elements
    saxpy << <(N + 255) / 256, 256 >> > (N, 2.0f, d_x, d_y);//Launching a Kernel;
    // the number of thread blocks in the grid；the number of threads in a thread block.
    //Thread blocks and grids can be made one-, two- or three-dimensional by passing dim3
    cudaMemcpy(y, d_y, N * sizeof(float), cudaMemcpyDeviceToHost);//get the results back to the host

    float maxError = 0.0f;
    for (int i = 0; i < N; i++)
        maxError = max(maxError, abs(y[i] - 4.0f));
    printf("Max error: %f\n", maxError);

    cudaFree(d_x);//Cleaning Up
    cudaFree(d_y);
    free(x);
    free(y);
}