一、Enhancing Memory AllocationEnhancing Memory Allocation优化内存分配或改进内存分配、增强内存分配。在前面的文章对CUDA内存管理和优化以及优化的方法“内存预取”进行了分析说明。本文将对CUDA中整体的内存优化进行分析并根据具体的层次说明与之相关的内存优化方法。在此基础之上协调处理每个层次的内存优化以期达到整体的最大优化的可能。二、影响性能的限制条件一般来说影响软件效率和性能的在排除硬件的硬限制后就是看软件的各层间的性能与接口传递的性能了。从开发者角度来看在应用环境已经固定的情况下运行核心CPU或GPU是直接影响运行速度和效率的根本条件。但在这些条件已经达到指标要求后就要考虑处理过程中对数据管理操作的间接问题了。而在这其中内存的管理和应用则是最首要的限制条件了。如果因为内存显存的分配管理出现问题则有可能产生无谓的效率损失即出现了内存可优化或增强的前提。在CUDA编程中内存相关优化或增强分为三层主机端内存管理有过操作系统知识的都明白内存的不足是硬件资源永远的痛。虚拟内存的出现以及相关的内存交换技术等本质都是源于内存资源的紧张。特别是内缓存命中率失效后导致的数据重新加载会极大的降低效率和处理的性能。所以如果能够在主机端实现内存的锁定让关键的内存应用无法被交换到硬盘或者说提高内存的命中率再辅以其它技术如DMA等则可能在整体上大幅提高数据处理的速度从而提高性能和效率。尤其是IO overlapped技术的应用更是能够明显的提高资源的利用效率但也显著的提高了编程的复杂度。在CUDA的发展过程中从早期的cudaHostAlloc()、cudaMallocHost()到CUDA11.2后的cudaMemcpyAsync()系列都可以实现内存物理页的锁定。特别是CUDA11.2之后提供了Stream-Ordered Memory Allocator流顺序内存分配器即通过cudaMemcpyAsync()等实现了内存分配与CUDA流的深度融合。GPU的内存优化重用有过大规模软件开发经验的都知道内存池技术在很多情况下往往是解决内存应用的重要手段。内存的分配和回收是一项复杂和耗费时间较长的工作。而如果只是把内存分配起来而不断的重用的话就大幅降低了内存分配和回收占用的系统资源。所以在GPU中同样也提供了内存池化的技术如RAPIDS内存管理器 (RMM)内存池化的实现有多种实现方式一般来说可以通过一次分配大内存然后再反复应用是一种比较常见的方式。它除了具备常规的内存池的优势如减少内存碎片还提高了Host和Device的异步应用。当然池化也其劣势典型就是对待复杂多变的内存分配变化时可能导致内存的浪费。而CUDA中提供的统一内存及其内存的预取进一步提高了内存使用的效率。这个在前面已经详述过此处不再重复。统一内存与上面的池化结合则进一步发挥了两者的优势。同样同统一内存中的预取一样内存建议Memory Advise也提供了UM的托管编程的可控能力并在长期的内存使用中提高了性能。池化的内存还提供了流感知的分配即流可以和内存池进行绑定以其使流的管理与内存的管理进行同步。新技术的应用硬件技术在不断的前进而相应的软件平台也会跟着变化。不论是从固件的指令还是到软件的操作接口环境的改变必然导致内存管理技术的更迭。在新的NVIDIA显卡出现后如H200等就必然会导致内存管理技术的相应调整变化甚至可能是革命性的变化。传统的NUMA内存管理技术通常是把GPU作为一个节点内存暴露给系统平台管理这极有可能屋CPU与GPU间的内存数据的动态交换有可能引发性能的不稳定。而CUDA也因此推出了CDMM模式的GPU内存管理即Coherent Driver-Based Memory Management基于驱动程序的连贯内存管理。它针对的主要是Grace Hopper/Blackwell架构的硬件平台。目的就是不再直接将GPU的显存暴露给系统平台而是通过驱动程序来管理GPU显存从而提高内存管理的粒度控制进而提高内存应用的性能。特别是在容器化的班干部CDMM模式有着更强的适应性。AI技术的迭代速度不断在加快而作为软件底层建筑之一的CUDA必然也要跟上硬件和上层应用的发展技术体系就是在传统与进步中不断的融合从量变到质变影响着整个应用系统的前进。三、内存优化的整体技术及CUDA中的应用在分析了内存对CUDA应用的影响和限制后可以看一看内存优化的技术及CUDA中有哪些具体的技术内存池内存池化意味着内存的分配和回收的性能损失和瓶颈得到有效的控制池化的内存往往对内存的重用起到了重要的作用并显著降低了内存碎片。这也是为什么在软件应用层、中间层、库、底层和操作系统中都被广泛应用的原因。而前面提到的Stream-Ordered Memory Allocator技术则显式的提供了池化的操作特与流的感知绑定。零拷贝和显存直接操作学习和分析网络高并发和数据库的底层读写操作时对Direct IO技术和DMA技术进行了反复的分析说明。同样它们对于显存的操作也有着相同的效果。GDSGPU Direct Storage就是一种发展的方向内存预取、建议和写合并内存Write-Combined这个就更好理解预取是为了减少数据加载等待的时间建议是为了让内存根据当前场景提高内存的应用效率和性能写合并则是减少数据对内存写入操作的次数。这都是很传统的增强或优化内存使用的方法虚拟内存管理这种虚拟内存管理不只是传统的虚拟内存它更倾向于管理。即通过虚拟内存的机制实现CUDA内存管理的更灵活和高效。它对内存的精细化管理提供了更好的支持特别是对内存池更加友好。在大内存的分配管理时更有优势。由于虚拟内存地址空间与物理显存地址空间的解耦让二者间的地址映射包括多级映射、地址的分段管理、内存共享等都变得更加方便快捷。合并应用这种属于一种比较典型的处理机制比如在C的智能指针中提供了自定义的删除器这个删除器可以支持直接使用C内存的回收机制。CUDA中也提供了类似的机制但是否支持的多优秀还需要看看再说Enhancing Memory Allocation对于CUDA来说是一个不断发展演进的方向它从最初的简单、粗暴的管理到不断的支持流感知、异步处理和内存池化技术等不断的适配新的软、硬件平台和库。有机融合各种技术提高GPU的显存利用率和性能。四、例程下面看一个虚拟内存管理的例程#includecuda_runtime.h#includedevice_launch_parameters.h#includestdio.h#includecuda.h#includestdlib.hintmain(){cuInit(0);CUdevice cuDev;cuDeviceGet(cuDev,0);CUcontext cuCtx;cuCtxCreate(cuCtx,0,cuDev);// query allocation granularityCUmemAllocationProp prop{};prop.typeCU_MEM_ALLOCATION_TYPE_PINNED;// fix memory allocation typeprop.location.typeCU_MEM_LOCATION_TYPE_DEVICE;// the memory will be allocated on the cuDevprop.location.idcuDev;// specify the cuDev for allocationsize_tgran;cuMemGetAllocationGranularity(gran,prop,CU_MEM_ALLOC_GRANULARITY_RECOMMENDED);printf(alloc granularity: %zu bytes\n,gran);// set allocation size (must be a multiple of the granularity)size_tallocSize1024*1024;// 1 MBallocSize(allocSizegran-1)/gran*gran;printf(alloc size (aligned): %zu bytes\n,allocSize);CUmemGenericAllocationHandle cuMemHandle;cuMemCreate(cuMemHandle,allocSize,prop,0);CUdeviceptr cuDevPtr;cuMemAddressReserve(cuDevPtr,allocSize,0,0,0);printf(Reserved virtual address: 0x%llx\n,(unsignedlonglong)cuDevPtr);// pyhsically map the allocated memory to the reserved virtual address rangecuMemMap(cuDevPtr,allocSize,0,cuMemHandle,0);// set access permissions for the allocated memory: allow read/write access from the cuDevCUmemAccessDesc cuAccesDesc{};cuAccesDesc.location.typeCU_MEM_LOCATION_TYPE_DEVICE;cuAccesDesc.location.idcuDev;cuAccesDesc.flagsCU_MEM_ACCESS_FLAGS_PROT_READWRITE;// read/write accesscuMemSetAccess(cuDevPtr,allocSize,cuAccesDesc,1);// use the allocated memory - set it to a specific value using cuMemsetD8cuMemsetD8(cuDevPtr,0xCB,allocSize);// copy and checkunsignedchar*hostBuf(unsignedchar*)malloc(allocSize);cuMemcpyDtoH(hostBuf,cuDevPtr,allocSize);// display the first few bytes to verifyprintf(display mem content after memset: );for(inti0;i16;i){printf(%02x ,hostBuf[i]);}printf(\n);// clean upfree(hostBuf);//umapcuMemUnmap(cuDevPtr,allocSize);// free virtual addresscuMemAddressFree(cuDevPtr,allocSize);// free physical memory allocationcuMemRelease(cuMemHandle);// destroy cuCtxcuCtxDestroy(cuCtx);return0;}上面的代码简单的实现了物理内存分配和虚拟地址控制以及映射、操作内存的方法如果哪一个API不太清晰可直接查看NIVIDIA的官网的相关接口描述。说明需要在VS2022中的“项目”-“属性”-“链接器”-“输入”的“附加依赖项”中增加“cuda.lib”的库的引入。五、总结通过上述的分析可以看到一般情况下解决问题的思想往往是相通的。这也是抽象指导实践的一个验证方法。显存的处理与传统的内存管理和优化技术在指导思想和原则没有什么不同。优秀的设计者往往是在掌握传统的思想后在具体的平台上有着独到的发展和应用。而实践的过程又反过来促进传统思想的进步和发展形成有机的统一共同推进技术整体的演进