CUDA Cpp并行计算二维和三维网格

NVIDIA 的 CUDA 是一种通用并行计算平台和编程模型，可利用 GPU 的并行处理能力来加速深度学习和其他计算密集型应用程序。CUDA 和 NVIDIA GPU 已被应用于许多需要高浮点计算性能的领域，包括：

计算金融
气候、天气和海洋建模
数据科学和分析
深度学习和机器学习
国防和情报
制造/AEC（建筑、工程和施工）：CAD 和 CAE（包括计算流体动力学、计算结构力学、设计和可视化以及电子设计自动化）
媒体和娱乐（包括动画、建模和渲染；色彩校正和颗粒管理；合成；整理和效果；编辑；编码和数字分发；直播图形；现场、审查和立体工具；以及天气图形）
医学影像
油和气
研究：高等教育和超级计算（包括计算化学和生物学、数值分析、物理学和科学可视化）
安全保障
工具和管理

CUDA C++ 并行编程

与任何程序一样，执行必须从 CPU 开始。我们将创建测试数据，在 CPU（主机）上处理它，计时并将数据传输到 GPU（设备），再次处理并将结果复制回主机。下面的代码使用 nvcc -std=c 11 cuda_multiply.cu 进行编译，并使用 nvprof 进行分析。

#include <chrono>
#include <iostream>__global__ void _cuda_parallel_multiplication(int count, int* test_data, int magnitude);int main() {int count = 60000000;    // 60 million elementsint* test_data = new int[count];for(int i = 0; i < count; i++)test_data[i] = i;// Perform calculation on host CPUauto t1 = std::chrono::high_resolution_clock::now();for(int i = 0; i < count; i++)test_data[i] = test_data[i] * 5;auto t2 = std::chrono::high_resolution_clock::now();// Copy data to deviceint* d_test_data;cudaMalloc(&d_test_data, count * sizeof(int));cudaMemcpy(d_test_data, test_data, count * sizeof(int), cudaMemcpyHostToDevice);// Copy results back to devicecudaDeviceSynchronize();cudaMemcpy(test_data, d_test_data, count * sizeof(int), cudaMemcpyDeviceToHost);cudaFree(d_test_data);for(int i = 0; i < 10; i++)std::cout << i << ": " << test_data[i] << std::endl;std::cout	<< "CPU time: "<< std::chrono::duration_cast<std::chrono::milliseconds>(t2 - t1).count()<< "ms" << std::endl;
}__global__ void _cuda_parallel_multiplication(int count, int* test_data, int magnitude) {int globalIdx = blockIdx.x * blockDim.x + threadIdx.x;while (globalIdx < count) {test_data[globalIdx] = test_data[globalIdx] * magnitude;globalIdx +=  blockDim.x * gridDim.x;__syncthreads();}
}

我们首先生成6000万个整数作为测试数据：

int main() {int count = 60000000;    // 60 million elementsint* test_data = new int[count];
for(int i = 0; i < count; i++)test_data[i] = i;
}

CPU 乘法，用 C++ chrono 计时，#include

auto t1 = std::chrono::high_resolution_clock::now();
for(int i = 0; i < count; i++)test_data[i] = test_data[i] * 5;
auto t2 = std::chrono::high_resolution_clock::now();

为了在GPU上执行计算，必须首先复制数据，CUDA运行时暴露cudaMalloc在GPU上分配内存，并暴露cudaMemcpy来传输数据。

int* d_test_data;
cudaMalloc(&d_test_data, count * sizeof(int));
cudaMemcpy(d_test_data, test_data, count * sizeof(int), cudaMemcpyHostToDevice);

…

CUDA 并行模型

从串行 CUDA 计算到并行 CUDA 计算的变化涉及硬件和软件的变化。硬件变化涉及包含多个计算单元以及用于调度计算和数据传输的机制的芯片。软件更改涉及 API 和编程语言的扩展。

GPU 实现并行化的最重要属性是该设备包含的不是一个或多个计算单元（如现代多核 CPU），而是数百或数千个计算单元。如果您可以将计算组织为大量独立的子任务，那么众多的计算单元使您能够并行执行其中许多任务；也就是说，同时执行任务而不是顺序执行任务。实现这种并行化涉及一些主要的组织问题：特定的计算单元如何知道要执行哪个子任务？大量计算单元如何访问它们所需的指令和数据而不造成严重的通信流量拥堵？

CUDA 采用单指令多线程 (SIMT) 并行化模型。 CUDA GPU 包含许多称为核心的基本计算单元，每个核心都包含算术逻辑单元 (ALU) 和浮点单元 (FPU)。核心被分为称为流式多处理器 (SM) 的组。

我们通过将计算任务分解为许多称为线程的子任务来并行化计算任务，这些子任务被组织成块。块被划分为 warp，其大小与 SM 中的核心数量相匹配。每个扭曲被分配给一个特定的 SM 来执行。 SM 的控制单元指示其每个核心为分配的 warp 中的每个线程同时执行相同的指令，因此称为“单指令多线程”。

执行相同的指令不仅仅是一种冗余练习，因为每个线程使用 CUDA提供的唯一索引值执行不同的计算。 SIMT 方法具有可扩展性，因为可以通过提供更多 SM 来分担计算负载来提高计算吞吐量。CPU 和 GPU 架构之间的对比， CPU 有几个计算核心，占据芯片的一小部分，并有很大的区域专门用于控制和缓存，以帮助这几个核心快速运行。一般规则是，访问数据所需的时间随着计算核心与存储数据的内存位置之间的距离的增加而增加。内核等待数据的时间称为延迟，CPU 的设计目的是通过专用大量空间来存储可以快速访问的数据来最大程度地减少延迟。

GPU上的空间分配有很大不同。大部分芯片区域专用于组织成具有共享控制单元的 SM 的大量计算核心。 GPU 的目标不是最小化延迟（这需要每个核心大量的缓存），而是隐藏延迟。当执行一个warp需要不可用的数据时，SM切换到执行另一个数据可用的warp。重点是整体计算吞吐量，而不是单个核心的速度。