做网站哈尔滨,无锡做网页,大型网站开发视频,三 网站开发使用软件环境以下是针对 2.2.3 GPU Device Memory Spaces 章节内容的详细知识点整理#xff0c;包含所有内存类型的特性、使用方法及最佳实践。2.2.3. GPU Device Memory Spaces (GPU设备内存空间) 内存类型概览表内存类型作用域生命周期物理位置关键特性全局内存网格#xff08;所有线程…以下是针对2.2.3 GPU Device Memory Spaces章节内容的详细知识点整理包含所有内存类型的特性、使用方法及最佳实践。2.2.3. GPU Device Memory Spaces (GPU设备内存空间)内存类型概览表内存类型作用域生命周期物理位置关键特性全局内存网格所有线程应用程序设备容量大延迟高所有线程可访问常量内存网格所有线程应用程序设备只读缓存优化延迟低共享内存块块内线程内核SM用户管理低延迟高带宽局部内存线程内核设备逻辑线程局部物理在全局内存寄存器线程内核SM最快存储编译器管理2.2.3.1. Global Memory (全局内存)基本概念别名设备内存device memory类比相当于CPU系统中的RAM作用内核中所有线程都可以访问的主要存储空间关键特性特性说明可访问性网格内的所有线程都可读/写持久性分配后持续存在直到显式释放或程序终止生命周期从分配到cudaFree()或cudaDeviceReset()分配与释放API操作API说明分配cudaMalloc()分配设备内存分配统一内存cudaMallocManaged()分配可由CPU/GPU访问的内存拷贝cudaMemcpy()主机与设备间数据传输释放cudaFree()释放设备内存使用流程// 1. 分配全局内存cudaMalloc(d_data,size);// 2. 初始化数据通过拷贝cudaMemcpy(d_data,h_data,size,cudaMemcpyHostToDevice);// 3. 内核中使用kernelgrid,block(d_data);// 4. 将结果拷贝回主机cudaMemcpy(h_result,d_data,size,cudaMemcpyDeviceToHost);// 5. 释放内存cudaFree(d_data);注意事项数据竞争多个线程同时读写同一位置需同步返回值内核是void类型只能通过全局内存将结果返回主机持久性数据在内核执行间持续存在可被多个内核使用示例vecAdd内核__global__voidvecAdd(float*A,float*B,float*C,intvectorLength){intworkIndexthreadIdx.xblockIdx.x*blockDim.x;if(workIndexvectorLength){C[workIndex]A[workIndex]B[workIndex];}}A、B、C指针指向全局内存所有线程可同时读取A、B写入C2.2.3.2. Shared Memory (共享内存)基本概念物理位置每个SM内部资源关系与L1缓存共享同一物理资源类比用户管理的暂存器scratchpad关键特性特性说明可访问性同一线程块内的所有线程生命周期内核执行期间性能比全局内存带宽更高、延迟更低容量较小依GPU架构而异同步机制__syncthreads()__syncthreads();作用阻塞块内所有线程直到全部到达该调用点用途确保所有线程对共享内存的写入完成后再读取注意在条件分支中使用需谨慎可能导致死锁同步示例__global__voidexample_syncthreads(int*input_data,int*output_data){__shared__intshared_data[128];// 每个线程写入共享内存的不同位置shared_data[threadIdx.x]input_data[threadIdx.x];// 同步确保所有写入完成__syncthreads();// 单个线程安全地读取所有共享数据if(threadIdx.x0){intsum0;for(inti0;iblockDim.x;i){sumshared_data[i];}output_data[blockIdx.x]sum;}}共享内存大小查询属性说明sharedMemPerMultiprocessor每个SM的共享内存总量sharedMemPerBlock每个线程块可用的最大共享内存cudaDeviceProp prop;cudaGetDeviceProperties(prop,device);printf(Shared memory per SM: %zu\n,prop.sharedMemPerMultiprocessor);printf(Shared memory per block: %zu\n,prop.sharedMemPerBlock);缓存配置cudaFuncSetCacheConfig()cudaFuncSetCacheConfig(kernel,cudaFuncCachePreferShared);// 选项// cudaFuncCachePreferNone - 无偏好// cudaFuncCachePreferShared - 偏好更多共享内存// cudaFuncCachePreferL1 - 偏好更多L1缓存// cudaFuncCachePreferEqual - 偏好平衡注意这只是给运行时的提示不保证一定采用运行时根据资源和内核需求自由决定2.2.3.2.1. Static Allocation (静态分配)语法__shared__floatsharedArray[1024];特点在内核内部声明使用__shared__说明符大小必须在编译时确定所有块内线程都可访问示例__global__voidkernel(){__shared__intcache[256];cache[threadIdx.x]threadIdx.x;__syncthreads();// 使用cache...}2.2.3.2.2. Dynamic Allocation (动态分配)启动时指定大小// 第三个参数指定共享内存大小字节kernelgrid,block,sharedMemoryBytes();内核内声明extern__shared__floatsharedArray[];多个动态数组的手动分区正确方法注意对齐extern__shared__floatarray[];short*array0(short*)array;// 2字节对齐float*array1(float*)array0[128];// 4字节对齐int*array2(int*)array1[64];// 4字节对齐错误示例未对齐extern__shared__floatarray[];short*array0(short*)array;float*array1(float*)array0[127];// ❌ 未4字节对齐对齐要求指针必须按指向类型对齐short需要2字节对齐float/int需要4字节对齐2.2.3.3. Registers (寄存器)基本概念物理位置SM内部作用域线程局部管理方式由编译器自动管理关键特性特性说明速度最快的存储类型容量有限依GPU架构而异分配编译器为每个线程分配查询regsPerMultiprocessor和regsPerBlock寄存器查询cudaDeviceProp prop;cudaGetDeviceProperties(prop,device);printf(Registers per SM: %d\n,prop.regsPerMultiprocessor);printf(Registers per block: %d\n,prop.regsPerBlock);控制寄存器使用-maxrregcountnvcc-maxrregcount32kernel.cu# 限制内核最多使用32个寄存器/线程影响减少寄存器使用→ 可在SM上调度更多线程块过度减少→ 导致寄存器溢出spilling变量被移至局部内存权衡并行度 vs 每个线程的性能2.2.3.4. Local Memory (局部内存)基本概念逻辑作用域线程局部物理位置全局内存中命名来源名称来自逻辑作用域而非物理位置使用场景编译器将自动变量放入局部内存的情况场景说明索引不确定的数组数组索引不是编译时常量大型结构或数组占用寄存器空间过大寄存器溢出内核使用寄存器超过可用数量性能特性特性说明延迟与全局内存相同高延迟带宽与全局内存相同合并访问连续线程访问连续32位字时可合并局部内存的合并访问局部内存组织方式连续32位字被连续线程ID访问只要warp内所有线程访问相同相对地址访问就是完全合并的示例// 如果arr被编译器放入局部内存intarr[4];arr[0]threadIdx.x;// 所有线程访问arr[0] → 合并访问arr[1]data[threadIdx.x];// 不同索引 → 可能无法合并内存类型对比总结特性寄存器局部内存共享内存全局内存常量内存位置SM内部设备内存SM内部设备内存设备内存作用域线程线程块网格网格生命周期内核内核内核应用程序应用程序速度最快慢快慢快缓存时管理编译器编译器程序员程序员程序员容量很小大小很大小但缓存最佳实践总结全局内存主要数据存储注意数据竞争合并访问以提高带宽利用率共享内存用于块内线程协作和频繁访问的数据使用__syncthreads()确保数据一致性动态分配时注意对齐要求寄存器编译器自动管理通常最优必要时通过-maxrregcount限制使用量避免过度限制导致寄存器溢出局部内存尽量使数组索引为编译时常量注意合并访问模式内存选择策略频繁访问的数据→ 共享内存只读数据→ 常量内存线程私有数据→ 寄存器大容量、跨线程共享数据→ 全局内存核心API总结表API用途cudaMalloc()分配全局内存cudaFree()释放全局内存cudaMemcpy()主机-设备数据传输cudaGetDeviceProperties()获取设备属性内存大小等cudaFuncSetCacheConfig()设置缓存/共享内存偏好__syncthreads()块内线程同步