почему cuda для передачи переменной устройства дает несмежную ошибку? - PullRequest
Новая
       38

почему cuda для передачи переменной устройства дает несмежную ошибку?

0 голосов
/ 16 июня 2020

Я новичок в cuda и пытаюсь вычислить матрицу с помощью gpu. Идея состоит в том, чтобы использовать индекс потока и индекс блока, чтобы найти разные участки памяти двух разных матриц. Вычислите матрицу поблочно, а затем скопируйте ее обратно в соответствующую выходную матрицу.

Код выглядит примерно так: 


import math
import numpy as np
from numba import cuda

@cuda.jit(device=True)
def GPU_SimpleOp(dim1, dim2):
    x = dim1 + dim2
    return x

@cuda.jit
def GPUCore(nd_matrix1, nd_matrix2, nd_output):

    blockix = cuda.blockIdx.x
    threadix = cuda.threadIdx.x

    if blockix > nd_matrix1.shape[0] or threadix > nd_matrix2.shape[0]:
        return

    nd_output[blockix, threadix] = GPU_SimpleOp(nd_matrix1[blockix], nd_matrix2[threadix])


def Test(nd_matrix1, nd_matrix2, blockno=256, threadno = 256):
    cuda.pinned(nd_matrix1)
    cuda.pinned(nd_matrix2)

    dim1 = nd_matrix1.shape[0]
    dim2 = nd_matrix2.shape[0]
    output = np.ndarray(shape=[dim1, dim2])
    nd_output = np.ascontiguousarray(output)
    cuda.pinned(nd_output)

    blocks_dim1 = int(math.ceil(dim1 / blockno))
    thread_dim2 = int(math.ceil(dim2 / threadno))

    for i in range(blocks_dim1):
        dim1_s = i
        dim1_e = min((i + 1) * blockno, nd_matrix1.shape[0])
        for j in range(thread_dim2):
            dim2_s = j
            dim2_e = min((j + 1) * threadno, nd_matrix2.shape[0])

            stream = cuda.stream()

            device_mat1 = cuda.to_device(nd_matrix1[dim1_s:dim1_e], stream)
            device_mat2 = cuda.to_device(nd_matrix2[dim2_s:dim2_e], stream)
            device_output = cuda.to_device(nd_output[dim1_s:dim1_e, dim2_s:dim2_e], stream)
            GPUCore[blockno, threadno, stream](device_mat1, device_mat2, device_output)

            device_output.copy_to_host(nd_output[dim1_s:dim1_e, dim2_s:dim2_e], stream)
            stream.synchronize()


if __name__ == '__main__':
    dim1 = 1000
    dim2 = 5000
    nd_matrix1 = np.random.random(dim1)
    nd_matrix2 = np.random.random(dim2)
    blockno = 256
    threadno = 256

    Test(nd_matrix1, nd_matrix2, blockno, threadno)

Когда я запускаю этот код, я получил сообщение об ошибке: 

Connected to pydev debugger (build 201.7846.77)
Traceback (most recent call last):
  File "D:\ProgramData\Anaconda3\lib\contextlib.py", line 130, in __exit__
    self.gen.throw(type, value, traceback)
  File "D:\ProgramData\Anaconda3\lib\site-packages\numba\cuda\cudadrv\devices.py", line 127, in ensure_context
    yield
  File "D:\ProgramData\Anaconda3\lib\site-packages\numba\cuda\cudadrv\devices.py", line 225, in _require_cuda_context
    return fn(*args, **kws)
  File "D:\ProgramData\Anaconda3\lib\site-packages\numba\cuda\api.py", line 110, in to_device
    to, new = devicearray.auto_device(obj, stream=stream, copy=copy)
  File "D:\ProgramData\Anaconda3\lib\site-packages\numba\cuda\cudadrv\devicearray.py", line 692, in auto_device
    sentry_contiguous(obj)
  File "D:\ProgramData\Anaconda3\lib\site-packages\numba\cuda\cudadrv\devicearray.py", line 666, in sentry_contiguous
    raise ValueError(errmsg_contiguous_buffer)
ValueError: Array contains non-contiguous buffer and cannot be transferred as a single memory region. Please ensure contiguous buffer with numpy .ascontiguousarray()

Ошибка кода в строке device_output = cuda.to_device(nd_output[dim1_s:dim1_e, dim2_s:dim2_e], stream). Хотя я погуглил ответ и попытался использовать nd_output = np.ascontiguousarray(output). Это все еще не работает. Любая помощь, пожалуйста?

1 Ответ

1 голос
/ 16 июня 2020

Я внес изменения в код в соответствии с предложениями талонмиев. 

import math
import numpy as np
from numba import cuda

@cuda.jit(device=True)
def GPU_SimpleOp(dim1, dim2):
    x = dim1 + dim2
    return x

@cuda.jit
def GPUCore(nd_matrix1, nd_matrix2, nd_output):

    blockix = cuda.blockIdx.x
    threadix = cuda.threadIdx.x

    if blockix > nd_matrix1.shape[0] or threadix > nd_matrix2.shape[0]:
        return

    nd_output[blockix, threadix] = GPU_SimpleOp(nd_matrix1[blockix], nd_matrix2[threadix])


def Test(nd_matrix1, nd_matrix2, blockno=256, threadno = 256):
    cuda.pinned(nd_matrix1)
    cuda.pinned(nd_matrix2)

    dim1 = nd_matrix1.shape[0]
    dim2 = nd_matrix2.shape[0]
    output = np.ndarray(shape=[dim1, dim2])
    nd_output = np.ascontiguousarray(output)
    cuda.pinned(nd_output)

    blocks_dim1 = int(math.ceil(dim1 / blockno))
    thread_dim2 = int(math.ceil(dim2 / threadno))

    for i in range(blocks_dim1):
        dim1_s = i * blockno
        dim1_e = min((i + 1) * blockno, nd_matrix1.shape[0])
        for j in range(thread_dim2):
            dim2_s = j * blockno
            dim2_e = min((j + 1) * threadno, nd_matrix2.shape[0])

            stream = cuda.stream()
            device_mat1 = cuda.to_device(nd_matrix1[dim1_s:dim1_e], stream)
            device_mat2 = cuda.to_device(nd_matrix2[dim2_s:dim2_e], stream)
            new_array = np.zeros_like(nd_output[dim1_s:dim1_e, dim2_s:dim2_e])
            device_output = cuda.to_device(new_array, stream)
            GPUCore[blockno, threadno, stream](device_mat1, device_mat2, device_output)

            device_output.copy_to_host(new_array, stream)
            nd_output[dim1_s:dim1_e, dim2_s:dim2_e] = new_array
            stream.synchronize()

if __name__ == '__main__':
    dim1 = 1000
    dim2 = 5000
    nd_matrix1 = np.random.random(dim1)
    nd_matrix2 = np.random.random(dim2)
    blockno = 256
    threadno = 256

    Test(nd_matrix1, nd_matrix2, blockno, threadno)

Теперь работает. Но меня это все равно смущает. Почему device_mat1 = cuda.to_device(nd_matrix1[dim1_s:dim1_e], stream) работает, а device_output = cuda.to_device(nd_output[dim1_s:dim1_e, dim2_s:dim2_e], stream) не работает?

...