我正在使用pyCUDA进行CUDA编程。我需要在内核函数中使用随机数。在它内部不工作(pyCUDA)。由于GPU中有大量的工作要做,在CPU内部产生随机数,然后再将它们传输到GPU上是行不通的,反而消除了使用GPU的动机。
补充质询:
发布于 2017-09-12 08:18:02
尽管您在问题中断言了什么,PyCUDA仍然对CUrand有相当全面的支持。GPUArray模块有一个直接的接口来使用主机端API来填充设备内存(请注意,在这种情况下,随机生成器在GPU上运行)。
在CUrand内核代码中使用来自PyCUDA的设备端API也是完全可能的。在这个用例中,最棘手的部分是为线程生成器状态分配内存。有三种选择--静态地在代码中,动态地使用主机内存侧分配,以及动态地使用设备端内存分配。下面的示例(经过了很小的测试)说明了后者,正如您在问题中询问的那样:
import numpy as np
import pycuda.autoinit
from pycuda.compiler import SourceModule
from pycuda import gpuarray
code = """
#include <curand_kernel.h>
const int nstates = %(NGENERATORS)s;
__device__ curandState_t* states[nstates];
__global__ void initkernel(int seed)
{
int tidx = threadIdx.x + blockIdx.x * blockDim.x;
if (tidx < nstates) {
curandState_t* s = new curandState_t;
if (s != 0) {
curand_init(seed, tidx, 0, s);
}
states[tidx] = s;
}
}
__global__ void randfillkernel(float *values, int N)
{
int tidx = threadIdx.x + blockIdx.x * blockDim.x;
if (tidx < nstates) {
curandState_t s = *states[tidx];
for(int i=tidx; i < N; i += blockDim.x * gridDim.x) {
values[i] = curand_uniform(&s);
}
*states[tidx] = s;
}
}
"""
N = 1024
mod = SourceModule(code % { "NGENERATORS" : N }, no_extern_c=True, arch="sm_52")
init_func = mod.get_function("_Z10initkerneli")
fill_func = mod.get_function("_Z14randfillkernelPfi")
seed = np.int32(123456789)
nvalues = 10 * N
init_func(seed, block=(N,1,1), grid=(1,1,1))
gdata = gpuarray.zeros(nvalues, dtype=np.float32)
fill_func(gdata, np.int32(nvalues), block=(N,1,1), grid=(1,1,1))
这里有一个初始化内核,需要运行一次才能为生成器状态分配内存并使用种子初始化它们,然后是一个使用这些状态的内核。如果要运行大量线程,则需要注意malloc堆大小限制,但可以通过PyCUDA驱动程序API接口对这些线程进行操作。
发布于 2019-03-14 10:55:18
我在接受的答案中有一个问题。我们有一个名叫mangling,有点恶心(这些_Z10initkerneli
和_Z14randfillkernelPfi
)。为了避免这种情况,我们可以手动将代码包装在extern "C" {...}
子句中。
code = """
#include <curand_kernel.h>
const int nstates = %(NGENERATORS)s;
__device__ curandState_t* states[nstates];
extern "C" {
__global__ void initkernel(int seed)
{ .... }
__global__ void randfillkernel(float *values, int N)
{ .... }
}
"""
然后,代码仍然用no_extern_c=True
编译。
mod = SourceModule(code % { "NGENERATORS" : N }, no_extern_c=True)
这应该适用于
init_func = mod.get_function("initkernel")
fill_func = mod.get_function("randfillkernel")
希望这能有所帮助。
https://stackoverflow.com/questions/46169633
复制相似问题