GPU Coder™为您提供两种不同的内存分配(malloc
)可在CUDA中使用的模式®编程模型,cudaMalloc
而且cudaMallocManaged
.cudaMalloc
API适用于传统的CPU分离,GPU全局内存。cudaMallocManaged
适用于统一内存.
从程序员的角度来看,传统的计算机架构要求在CPU和GPU内存空间之间分配和共享数据。应用程序需要管理这两个内存空间之间的数据传输,这增加了复杂性。统一内存创建一个托管内存池,由CPU和GPU共享。CPU和GPU都可以通过一个指针访问托管内存。统一内存试图通过将数据迁移到需要它的设备来优化内存性能,同时对程序隐藏迁移细节。尽管统一内存简化了编程模型,但当GPU上写入的数据在CPU上被访问时,它需要设备同步调用。GPU Coder插入这些同步调用。根据NVIDIA®当使用CUDA 8.0时,统一内存可以提供显著的性能优势,或者当针对嵌入式硬件时,如NVIDIA Tegra®.
若要更改GPU Coder应用程序中的内存分配模式,请使用Malloc模式
下拉框更多设置->GPU编码器.在使用命令行界面时,请使用MallocMode
生成配置属性并将其设置为任意一个“离散”
或“统一”
.
GPU Coder分析CPU和GPU分区之间的数据依赖关系,并进行优化,使CPU和GPU分区之间的数据依赖关系最小化cudaMemcpy
生成的代码中的函数调用。分析还确定了数据必须在CPU和GPU之间复制的最小位置集cudaMemcpy
.
例如,函数喷火
具有在CPU上按顺序处理数据和在GPU上并行处理数据的代码段。
function [out] = foo(input1,input2)…% CPU工作input1 =…input2 =…tmp1 =…tmp2 = ... ... % GPU工作内核1(gpuInput1, gpuTmp1);kernel2(gpuInput2, gpuTmp1, gpuTmp2);kernel3(gpuTmp1, gpuTmp2, gpuOut);…% CPU工作…=结束
一个未优化的CUDA实现可能有多个cudaMemcpy
函数调用来传递所有输入gpuInput1, gpuInput2
,和暂时的结果gpuTmp1, gpuTmp2
在内核调用之间。因为中间结果gpuTmp1, gpuTmp2
不使用GPU外部,它们可以存储在GPU内存内,从而减少cudaMemcpy
函数调用。这些优化提高了生成代码的整体性能。优化后的实现为:
gpuInput1 = input1;gpuInput2 = input2;kernel1<<< >>>(gpuInput1, gpuTmp1);kernel2<<< >>>(gpuInput2, gpuTmp1, gpuTmp2);kernel3<<< >>>(gpuTmp1, gpuTmp2, gpuOut);out = gpuOut;
消除冗余cudaMemcpy
调用,GPU Coder分析给定变量的所有使用和定义,并使用状态标志来执行最小化。这个表中显示了原始代码的示例以及生成的代码的外观。
原始代码 | 优化生成的代码 |
---|---|
A(:) = ... ... for i = 1:N gB = kernel1(gA);gA = kernel2(gB);if (somcondition) gC = kernel3(gA, gB);... ... = C; |
A(:) =…A_isDirtyOnCpu = true;i = 1:N if (A_isDirtyOnCpu) gA = A;A_isDirtyOnCpu = false;end gB = kernel1(gA);gA = kernel2(gB);if (somcondition) gC = kernel3(gA, gB);C_isDirtyOnGpu = true;if (C_isDirtyOnGpu) C = gC;C_isDirtyOnGpu = false;结束,结束; |
的_isDirtyOnCpu
flag告诉GPU Coder内存优化例程,给定变量被声明并在CPU或GPU上使用。