CUDA 编程基础

# 验证CUDA安装
nvcc --version          # CUDA编译器版本
nvidia-smi              # GPU信息和驱动版本
ls /usr/local/cuda/     # CUDA安装目录

# 编译CUDA程序
nvcc -o hello hello.cu
./hello

# 常用编译选项
nvcc -arch=sm_80 -O3 -o output input.cu  # 指定GPU架构

// Hello CUDA - 你的第一个GPU程序
#include <stdio.h>
#include <cuda_runtime.h>

__global__ void hello_from_gpu() {
    printf("Hello from GPU thread %d!\n", threadIdx.x);
}

int main() {
    // 启动1个Block，10个Thread
    hello_from_gpu<<<1, 10>>>();
    
    // 等待GPU完成
    cudaDeviceSynchronize();
    
    return 0;
}

// 多维线程索引
__global__ void matrix_2d(float *data, int width, int height) {
    // 计算2D索引
    int col = threadIdx.x + blockIdx.x * blockDim.x;
    int row = threadIdx.y + blockIdx.y * blockDim.y;
    
    if (col < width && row < height) {
        int idx = row * width + col;
        data[idx] = data[idx] * 2.0f;
    }
}

// 启动2D Grid
dim3 block(16, 16);  // 16x16 = 256 threads per block
dim3 grid((width+15)/16, (height+15)/16);
matrix_2d<<<grid, block>>>(d_data, width, height);

// 内存类型使用示例
// Constant Memory - 编译时初始化
__constant__ float d_weights[256];

// Shared Memory - Block内共享
__global__ void kernel_with_shared(float *input, float *output) {
    __shared__ float shared_data[256];
    
    // 协作加载数据到Shared Memory
    shared_data[threadIdx.x] = input[threadIdx.x + blockIdx.x * blockDim.x];
    __syncthreads();  // 同步所有线程
    
    // 使用Shared Memory计算
    output[threadIdx.x + blockIdx.x * blockDim.x] = 
        shared_data[threadIdx.x] * d_weights[threadIdx.x];
}

// 完整的内存管理示例
int main() {
    const int N = 1024;
    size_t size = N * sizeof(float);
    
    // Host内存
    float *h_a = (float*)malloc(size);
    float *h_b = (float*)malloc(size);
    
    // Device内存
    float *d_a, *d_b;
    cudaMalloc(&d_a, size);
    cudaMalloc(&d_b, size);
    
    // Host → Device
    cudaMemcpy(d_a, h_a, size, cudaMemcpyHostToDevice);
    
    // 启动Kernel
    my_kernel<<<N/256, 256>>>(d_a, d_b, N);
    
    // Device → Host
    cudaMemcpy(h_b, d_b, size, cudaMemcpyDeviceToHost);
    
    // 清理
    free(h_a); free(h_b);
    cudaFree(d_a); cudaFree(d_b);
}

// 专业级错误检查宏
#define CUDA_CHECK(call) \
    do { \
        cudaError_t err = call; \
        if (err != cudaSuccess) { \
            fprintf(stderr, "CUDA Error in %s at line %d: %s\n", \
                __FILE__, __LINE__, cudaGetErrorString(err)); \
            exit(EXIT_FAILURE); \
        } \
    } while(0)

// Kernel启动后检查
my_kernel<<<grid, block>>>(args);
CUDA_CHECK(cudaGetLastError());      // 检查启动错误
CUDA_CHECK(cudaDeviceSynchronize()); // 检查执行错误