解决 pyopencl 数组偏移限制

Posted 2023-02-16

【中文标题】解决 pyopencl 数组偏移限制

【英文标题】：Working around pyopencl array offset limitation

【发布时间】：2013-09-03 09:54:33

【问题描述】：

有没有办法解决 PyOpenCL 中的限制：

array.data

失败

pyopencl.array.ArrayHasOffsetError: 您正在尝试的操作尚不支持从缓冲区开头偏移开始的数组。

我试过了：

a.base_data[a.offset: a.offset + a.nbytes]

这似乎有时有效，但其他时候我得到：

pyopencl.LogicError: clCreateSubBuffer 失败：值无效

【参考方案1】：

clcreateSubBuffer 需要具有对齐的偏移量（或者在这种情况下称为origin），并且size + origin 需要在缓冲区的范围内。

如果指定区域，则在 errcode_ret 中返回 CL_INVALID_VALUE (origin, size) 超出缓冲区范围。

CL_MISALIGNED_SUB_BUFFER_OFFSET 在 errcode_ret 中返回 上下文中没有与源缓冲区关联的设备 值与 CL_DEVICE_MEM_BASE_ADDR_ALIGN 值对齐。

对于您看到的特定错误，您的程序或 pyopencl 错误地计算了偏移后数组的大小。即使您解决了这个问题，如果原始偏移量未与CL_DEVICE_MEM_BASE_ADDR_ALIGN 对齐，您可能仍然会遇到问题。

话虽如此，NVIDIA 似乎打破了规范并允许任意偏移。因此，您的里程可能会因硬件而异。

【讨论】：

感谢您的回答。不幸的是，我确实检查了0 <= a.offset < a.offset + a.nbytes <= a.base_data + a.size，所以我认为不是这样……这可能是 PyOpenCL 独有的问题吗？

@NeilG 也许。在其他情况下，可能会返回相同的错误。不看 pyopencl 代码，我不想进一步推测。

【参考方案2】：

如果您只是想获得一个标记数组数据开始的缓冲区，以传递给内核，您不必担心大小。这是一个获取指向偏移数据开始的大小为 1 的缓冲区的函数：

def data_ptr(array):
    if array.offset:
        return array.base_data.get_sub_region(array.offset, 1)
    else:
        return array.data

如果你需要一个指向偏移数据开始的指针，你可以使用它来传递给内核。这是一个示例，我想将数组clA 的子区域clV 设置为值3。我使用data_ptr 获取指向clV 数据开头的指针。

import numpy as np
import pyopencl as cl
import pyopencl.array
ctx = cl.create_some_context()
queue = cl.CommandQueue(ctx)

m, n = 5, 5
A = np.random.uniform(size=(m, n)).astype(np.float32)
clA = cl.array.Array(queue, A.shape, A.dtype)
clA.set(A)

clV = clA[1::2, 1::2]

def data(array):
    if array.offset:
        return array.base_data.get_sub_region(array.offset, 1)
    else:
        return array.data

source = """
__kernel void fn(long si, long sj, __global float *Y)

    const int i = get_global_id(0);
    const int j = get_global_id(1);
    Y[i*si + j*sj] = 3;

"""
kernel = cl.Program(ctx, source).build().fn
gsize = clV.shape
lsize = None

estrides = np.array(clV.strides) / clV.dtype.itemsize
kernel(queue, gsize, lsize, estrides[0], estrides[1], data_ptr(clV))

print(clA.get())

【讨论】：

我明白了，您将指针传递给 opencl 程序并让它进行偏移？我猜这行得通，但绝对不理想。

不，偏移是在程序之前完成的（即它不在内核内部）。

您将索引作为参数发送到数组中，不是吗？