PyCUDA LogicError: cuModuleLoadDataEx failed: an illegal memory access was encountered
PyCUDA LogicError: cuModuleLoadDataEx failed: an illegal memory access was encountered
我正在尝试将双调排序与 pycuda 并行化。为此，我使用 SourceModule 和并行双调排序的 C 代码。对于内存副本管理，我使用 pycuda.driver 的 InOut 来简化一些内存传输
import pycuda.autoinit
import pycuda.driver as drv
from pycuda.compiler import SourceModule
from pycuda import gpuarray
import numpy as np
from time import time
ker = SourceModule(
"""
__device__ void swap(int & a, int & b){
int tmp = a;
a = b;
b = tmp;
}
__global__ void bitonicSort(int * values, int N){
extern __shared__ int shared[];
int tid = threadIdx.x + blockDim.x * blockIdx.x;
// Copy input to shared mem.
shared[tid] = values[tid];
__syncthreads();
// Parallel bitonic sort.
for (int k = 2; k <= N; k *= 2){
// Bitonic merge:
for (int j = k / 2; j>0; j /= 2){
int ixj = tid ^ j;
if (ixj > tid){
if ((tid & k) == 0){
//Sort ascending
if (shared[tid] > shared[ixj]){
swap(shared[tid], shared[ixj]);
}
}
else{
//Sort descending
if (shared[tid] < shared[ixj]){
swap(shared[tid], shared[ixj]);
}
}
}
__syncthreads();
}
}
values[tid] = shared[tid];
}
"""
)
N = 8 #lenght of A
A = np.int32(np.random.randint(1, 20, N)) #random numbers in A
BLOCK_SIZE = 256
NUM_BLOCKS = (N + BLOCK_SIZE-1)//BLOCK_SIZE
bitonicSort = ker.get_function("bitonicSort")
t1 = time()
bitonicSort(drv.InOut(A), np.int32(N), block=(BLOCK_SIZE,1,1), grid=(NUM_BLOCKS,1), shared=4*N)
t2 = time()
print("Execution Time {0}".format(t2 - t1))
print(A)
在内核中我使用extern __shared__，在pycuda中我使用共享参数与相应的4*N。也尝试在内核中使用 __shared__ int shared[N] 但它也不起作用（检查此处：）
运行 在 Google 协作中我收到以下错误：
/usr/local/lib/python3.6/dist-packages/pycuda/compiler.py in __init__(self, source, nvcc, options, keep, no_extern_c, arch, code, cache_dir, include_dirs)
292
293 from pycuda.driver import module_from_buffer
--> 294 self.module = module_from_buffer(cubin)
295
296 self._bind_module()
LogicError: cuModuleLoadDataEx failed: an illegal memory access was encountered
有谁知道是什么导致了这个错误？
您的设备代码没有正确考虑数组的大小。
您将在单个块中启动 256 个线程。这意味着您将有 256 个线程，其中 tid 编号为 0..255，试图执行每一行代码。例如，在这种情况下：
shared[tid] = values[tid];
例如，您将有一个线程试图做 shared[255] = values[255];
您的 shared 和 values 数组都没有那么大。这就是非法内存访问错误的原因。
解决这类琐碎问题的最简单方法是让数组大小与块大小匹配。
BLOCK_SIZE = N
根据我的测试，该更改会清除所有错误并生成正确排序的数组。
它不适用于 N 大于 1024 或多块使用，但是您的代码无论如何都必须针对多块排序进行修改。
如果您在进行更改后仍然遇到问题，我建议您重新启动 python 会话或 colab 会话。
我正在尝试将双调排序与 pycuda 并行化。为此，我使用 SourceModule 和并行双调排序的 C 代码。对于内存副本管理，我使用 pycuda.driver 的 InOut 来简化一些内存传输
import pycuda.autoinit
import pycuda.driver as drv
from pycuda.compiler import SourceModule
from pycuda import gpuarray
import numpy as np
from time import time
ker = SourceModule(
"""
__device__ void swap(int & a, int & b){
int tmp = a;
a = b;
b = tmp;
}
__global__ void bitonicSort(int * values, int N){
extern __shared__ int shared[];
int tid = threadIdx.x + blockDim.x * blockIdx.x;
// Copy input to shared mem.
shared[tid] = values[tid];
__syncthreads();
// Parallel bitonic sort.
for (int k = 2; k <= N; k *= 2){
// Bitonic merge:
for (int j = k / 2; j>0; j /= 2){
int ixj = tid ^ j;
if (ixj > tid){
if ((tid & k) == 0){
//Sort ascending
if (shared[tid] > shared[ixj]){
swap(shared[tid], shared[ixj]);
}
}
else{
//Sort descending
if (shared[tid] < shared[ixj]){
swap(shared[tid], shared[ixj]);
}
}
}
__syncthreads();
}
}
values[tid] = shared[tid];
}
"""
)
N = 8 #lenght of A
A = np.int32(np.random.randint(1, 20, N)) #random numbers in A
BLOCK_SIZE = 256
NUM_BLOCKS = (N + BLOCK_SIZE-1)//BLOCK_SIZE
bitonicSort = ker.get_function("bitonicSort")
t1 = time()
bitonicSort(drv.InOut(A), np.int32(N), block=(BLOCK_SIZE,1,1), grid=(NUM_BLOCKS,1), shared=4*N)
t2 = time()
print("Execution Time {0}".format(t2 - t1))
print(A)
在内核中我使用extern __shared__，在pycuda中我使用共享参数与相应的4*N。也尝试在内核中使用 __shared__ int shared[N] 但它也不起作用（检查此处：
运行 在 Google 协作中我收到以下错误：
/usr/local/lib/python3.6/dist-packages/pycuda/compiler.py in __init__(self, source, nvcc, options, keep, no_extern_c, arch, code, cache_dir, include_dirs)
292
293 from pycuda.driver import module_from_buffer
--> 294 self.module = module_from_buffer(cubin)
295
296 self._bind_module()
LogicError: cuModuleLoadDataEx failed: an illegal memory access was encountered
有谁知道是什么导致了这个错误？
您的设备代码没有正确考虑数组的大小。
您将在单个块中启动 256 个线程。这意味着您将有 256 个线程，其中 tid 编号为 0..255，试图执行每一行代码。例如，在这种情况下：
shared[tid] = values[tid];
例如，您将有一个线程试图做 shared[255] = values[255];
您的 shared 和 values 数组都没有那么大。这就是非法内存访问错误的原因。
解决这类琐碎问题的最简单方法是让数组大小与块大小匹配。
BLOCK_SIZE = N
根据我的测试，该更改会清除所有错误并生成正确排序的数组。
它不适用于 N 大于 1024 或多块使用，但是您的代码无论如何都必须针对多块排序进行修改。
如果您在进行更改后仍然遇到问题，我建议您重新启动 python 会话或 colab 会话。