CUDA编程 5 执行流 和 运行库【参加CUDA线上训练营】

1. CUDA流

CUDA流是在CUDA并行编程模型中用于管理并行任务的单元。它提供了一种方法来组织多个Kernel的执行顺序，从而可以控制Kernel之间的同步。CUDA流是由多个线程组成的，线程以块的形式运行在一个计算单元上，每个线程执行一个Kernel。

一个CUDA流可以用于管理多个Kernel，其中每个Kernel可以在不同的线程块中并行运行。这样，开发者可以更灵活地设置执行顺序，同时管理多个Kernel的并行执行。此外，CUDA流还提供了同步机制，以确保Kernel之间的执行顺序。

使用CUDA流编写代码需要遵循一些原则，包括对线程块的组织，对共享存储器的管理，以及对错误的处理。在实际的开发过程中，开发者需要对程序进行优化，以确保最佳性能。

2. cuBLAS  

cuBLAS是CUDA平台上的一组高性能的BLAS（Basic Linear Algebra Subprograms）库。它提供了一组标准的线性代数运算，例如矩阵乘法、矩阵点乘和向量范数计算等，这些运算在计算机科学和工程领域中非常重要。

cuBLAS通过将算法转换为CUDA内核来加速运算，可以利用GPU的计算能力来实现高效的计算。它提供了一个简单易用的接口，使用者可以方便地调用cuBLAS库中的函数，而不需要关注底层的实现细节。

cuBLAS适用于许多不同的应用场景，例如图像处理、数据分析、机器学习等。此外，cuBLAS还支持多GPU环境，可以在多个GPU之间进行分布式计算，从而加速大规模的计算任务。

总的来说，cuBLAS是一个高效实用的CUDA平台上的线性代数库，是加速科学计算和工程应用的重要工具。

课后作业：

1. 尝试调用cublas做矩阵乘法和向量加法操作

#include <cuda_runtime.h>
#include <cublas_v2.h>
#include <iostream>

const int N = 3;

int main() {
    // Step 1: Initialize CUBLAS
    cublasHandle_t handle;
    cublasCreate(&handle);

    // Step 2: Allocate device memory
    float *A, *B, *C;
    cudaMalloc((void **)&A, N * N * sizeof(float));
    cudaMalloc((void **)&B, N * N * sizeof(float));
    cudaMalloc((void **)&C, N * N * sizeof(float));

    // Step 3: Copy data from host to device
    float h_A[N][N] = {{1, 2, 3}, {4, 5, 6}, {7, 8, 9}};
    float h_B[N][N] = {{10, 11, 12}, {13, 14, 15}, {16, 17, 18}};
    float h_C[N][N] = {{0, 0, 0}, {0, 0, 0}, {0, 0, 0}};
    cudaMemcpy(A, h_A, N * N * sizeof(float), cudaMemcpyHostToDevice);
    cudaMemcpy(B, h_B, N * N * sizeof(float), cudaMemcpyHostToDevice);

    // Step 4: Call CUBLAS to perform matrix multiplication
    const float alpha = 1.0f;
    const float beta = 0.0f;
    cublasSgemm(handle, CUBLAS_OP_N, CUBLAS_OP_N, N, N, N, &alpha, A, N, B, N, &beta, C, N);

    // Step 5: Copy result from device to host
    cudaMemcpy(h_C, C, N * N * sizeof(float), cudaMemcpyDeviceToHost);

    // Step 6: Print result
    for (int i = 0; i < N; ++i) {
        for (int j = 0; j < N; ++j) {
            std::cout << h_C[i][j] << " ";
        }
        std::cout << std::endl;
    }

    // Step 7: Clean up
    cudaFree(A);
    cudaFree(B);
    cudaFree(C);
    cublasDestroy(handle);
    return 0;
}

2. 如果本地文件存储着2个1000000*1000000的矩阵，我们想将这两个矩阵进行乘积，如何操作？

#include <iostream>
#include <cuda_runtime.h>
#include <cublas_v2.h>

#define M 1000000
#define N 1000000

__global__ void matrixMultiplicationKernel(float *A, float *B, float *C, int width) {
  int col = blockIdx.x * blockDim.x + threadIdx.x;
  int row = blockIdx.y * blockDim.y + threadIdx.y;
  
  if (col < width && row < width) {
    float value = 0;
    for (int i = 0; i < width; ++i) {
      value += A[row * width + i] * B[i * width + col];
    }
    C[row * width + col] = value;
  }
}

int main() {
  // 数据初始化
  float *h_A, *h_B, *h_C;
  h_A = new float[M * N];
  h_B = new float[M * N];
  h_C = new float[M * N];

  for (int i = 0; i < M * N; ++i) {
    h_A[i] = i;
    h_B[i] = i;
  }

  // 分配GPU内存并复制数据
  float *d_A, *d_B, *d_C;
  cudaMalloc((void**)&d_A, M * N * sizeof(float));
  cudaMalloc((void**)&d_B, M * N * sizeof(float));
  cudaMalloc((void**)&d_C, M * N * sizeof(float));

  cudaMemcpy(d_A, h_A, M * N * sizeof(float), cudaMemcpyHostToDevice);
  cudaMemcpy(d_B, h_B, M * N * sizeof(float), cudaMemcpyHostToDevice);

  // 设置线程数和块数
  dim3 blockDim(32, 32);
  dim3 gridDim((N + blockDim.x - 1) / blockDim.x, (N + blockDim.y - 1) / blockDim.y);
  
  // 执行矩阵乘法操作
  matrixMultiplicationKernel<<<gridDim, blockDim>>>(d_A, d_B, d_C, N);
  
  // 复制计算结果到主内存并释放GPU内存
  cudaMemcpy(h_C, d_C, M * N * sizeof(float), cudaMemcpyDeviceToHost);