CUDA 9.2中curand_init函数运行极慢问题咨询
问题分析与修复方案
这绝对是CUDA 9.2中curand库的行为变更导致的性能退化,不是你的代码bug——我之前也遇到过几乎一模一样的问题!
为什么会变慢?
CUDA 9.2对默认的XORWOW随机数生成器的curand_init实现做了调整,当大量线程使用相同的seed+不同的sequence参数初始化独立状态时,内部会触发大量的全局内存同步操作,导致初始化时间从毫秒级暴增到分钟级。而CUDA 9.1的实现没有这个同步开销,所以之前跑得飞快。
你的代码里刚好踩中了这个点:所有线程用同一个time(NULL)作为seed,用线程index作为sequence,完美触发了这个性能陷阱。
修复方案(两种可选)
方案1:修改seed的传递方式(最简单,无需改生成算法)
把每个线程的seed改成seed + index,sequence固定为0,这样每个线程的初始化参数都是唯一的,避开了同步开销:
修改后的setup_curand_state内核:
__global__ void setup_curand_state (curandState *state, int seed, int dim) { int index = threadIdx.x+blockDim.x*blockIdx.x; if (index < dim) // 用seed+index作为每个线程的独立seed,sequence设为0 curand_init(seed + index, 0, 0, &state[index]); }
方案2:切换到更快的MRG32k3a生成器
如果你不想改seed逻辑,可以改用MRG32k3a算法,它的初始化性能在CUDA 9.2中不受影响。需要用curand_init_ex指定生成器类型:
修改后的setup_curand_state内核:
__global__ void setup_curand_state (curandState *state, int seed, int dim) { int index = threadIdx.x+blockDim.x*blockIdx.x; if (index < dim) curand_init_ex(CURAND_RNG_PSEUDO_MRG32K3A, seed, index, 0, &state[index]); }
注意:使用这个算法时,后续的随机数生成函数(比如curand_normal)依然可以正常使用,curandState会自动适配算法类型。
额外小建议
你的代码里cudaDeviceSynchronize()只在setup之后调用,最好在set_random内核执行完后也加一个,避免提前释放内存导致的潜在问题——虽然当前代码可能没问题,但养成好习惯能减少后续踩坑的概率。
内容的提问来源于stack exchange,提问作者zmxnc
相关产品推荐
相关产品推荐

