如何在SYCL 2020 / DPC++中实现自定义4维数组查看器/ Package 器？

0md85ypi 于 2023-06-07 发布在其他

关注(0)|答案(3)|浏览(190)

在传统的C++中，可以通过（1）定义一个自定义的ArrayWrapper类，（2）重写()或[]运算符作为其“访问器”，在此成员函数中进行地址计算，以及（3）返回对值的引用，来创建线性内存中1D缓冲区的多维“查看器”或“ Package 器”。因此，4D数组可以通过语法糖array(a, b, c, d) = val访问。这提高了代码的可读性，并且还将查看器与阵列的实际存储器布局分离。

#include <iostream>
#include <cstdlib>

template <typename T>
class ArrayWrapper
{
public:
        ArrayWrapper(T *buf) : array(buf) {};

        inline T& operator() (size_t a, size_t b, size_t c, size_t d)
        {
                return array[a + b + c + d];
        }

        const inline T& operator() (size_t a, size_t b, size_t c, size_t d) const
        {
                return array[a + b + c + d];
        }
        T *array;
};

int main(void)
{
        int *buf = (int *) malloc(sizeof(int) * 100);
        ArrayWrapper<int> array(buf);
        array(1, 2, 3, 4) = 42;

        std::cout << array(1, 2, 3, 4) << std::endl;
}

但是，此 Package 器在DPC++ / SYCL 2020内核中不可用。

int main(void)
{
        sycl::queue Q;
        auto buf = sycl::malloc_shared<int>(20, Q);
        ArrayWrapper<int> array(buf);

        Q.single_task([=]() {
                array(1, 2, 3, 4) = 42;
        });
        Q.wait();

        std::cout << array(1, 2, 3, 4) << std::endl;
}

使用“英特尔DPC++”编译器编译此函数将返回以下错误：

question-sycl.cpp:37:21: error: expression is not assignable
                array(1, 2, 3, 4) = 42;
                ~~~~~~~~~~~~~~~~~ ^
1 error generated.
make: *** [Makefile:8: question-sycl.elf] Error 1

这是由于使用C++ lambda函数的结果，该函数默认情况下将其范围外的变量“捕获”为const变量。在传统的C中，这可以通过显式地要求lambda函数捕获引用Q.single_task([&array]() {}来解决，或者将lambda声明为可变函数Q.single_task([=]() mutable {}。但是，这两种用法在SYCL中似乎都不受支持，并且被DPC编译器禁止。
有没有办法在DPC++ / SYCL 2020中实现相同的语法糖array(a, b, c, d) = val？我注意到SYCL中的内存访问是由两个抽象提供的，分别称为缓冲区和访问器。不幸的是，它们仅支持1D、2D或3D阵列，而不支持更高维度。定义一个方便访问高维数组的 Package 器的最佳方法是什么？

c++

来源：https://stackoverflow.com/questions/76411362/how-do-i-implement-a-custom-4-dimensional-array-viewer-wrapper-in-sycl-2020-dp

3条答案

按热度按时间

fykwrbwg1#

正如你所说，SYCL中捕获的对象是不可变的，这是有充分理由的：目前还不清楚是所有的工作项都应该访问内核参数的共享对象，还是每个工作项都应该有自己的副本--最终，这在很大程度上取决于后端/硬件以及它们想要做什么。所以我们决定所有的SYCL内核参数都应该是不可变的。
你有两个选择（你已经找到了一个）：
1.只需复制内核参数：

Q.single_task([=]() {
   ArrayWrapper<int> a_kernel = array;
   a_kernel(1, 2, 3, 4) = 42;
 });

1.你已经找到了这个：如果这是不可接受的，请考虑 Package 器的常量模型。SYCL内核参数的不变性真正关心的是内核参数中的数据是否发生了变化。在您的例子中， Package 器只提供一个视图-- Package 器对象本身不会改变。因此，让const重载返回一个非常量引用可能是可以接受的，这将解决您的问题，并且可能更适合您的用例。请注意，真正的const视图仍然可以通过使用const T类型示例化 Package 器来表示。如果愿意，您甚至可以实现从ArrayWrapper<T>到ArrayWrapper<const T>的转换。正如您所说，这就是sycl::accessor对象的实现方式。
我想指出的是，您可能不必实现自己的高维数组 Package 器。您应该能够使用已经提供此功能的mdspan，并使用SYCL USM指针对其进行初始化。我不知道DPC++，但我知道这在hipSYCL / Open SYCL中工作。

赞(0）回复(0）举报 2023-06-07

rt4zxlrg2#

DPC++目前不支持可变闭包
解决方案是捕获指向array的指针：

auto f = [array = &array]() {
        (*array)(1, 2, 3, 4) = 42;
    };

sycl::queue Q对我来说不可用，因此我通过删除不相关的Q来简化示例。

赞(0）回复(0）举报 2023-06-07

llmtgqce3#

更新：这里的问题是SYCL内核期望array的operator()的const成员函数，但这里提供了两个定义，一个是返回引用的非常量成员函数，另一个是返回const值的常量成员函数。结果，const版本的函数被匹配，返回const值。根据定义，这是不能修改的。
因此，解决方案是删除operator()的const-return版本：

const inline T& operator() (size_t a, size_t b, size_t c, size_t d) const
{
        return array[a + b + c + d];
}

将成员函数的非const-return版本更改为const成员函数，如下所示：

inline T& operator() (size_t a, size_t b, size_t c, size_t d) const
{
    return array[a + b + c + d];
}

问题解决了
我注意到SYCL访问器本身也是这样实现的。
当将ArrayWrapper传递给lambda函数（计算内核）中的其他函数时，参数必须声明为const，而实际上它们是可修改的。这是一种令人困惑的用法，但在SYCL / DPC++编程中是一种预期的用法。因此，这些 Package 器不应该被称为Array，而是ArrayWrapper或ArrayAccessor，以突出只有 Package 器本身是const-数据不是。
过时的答案：273 K的回答给了我一个关于捕获指针的可能解决方案的提示，如：

Q.single_task([array = &array]() {
        (*ptr)(1, 2, 3, 4) = 42;
});
Q.wait();

不幸的是，它通过了编译器检查，但所有内存写入都没有效果，并且对主机不可见。我猜这是一个未定义的行为，它是由于对SYCL关于主机和设备之间共享内存的假设的微妙违反而引起的。
但是，在lambda函数中获取引用的地址是可行的：

Q.single_task([=]() {
        auto ptr = &array;
        (*ptr)(1, 2, 3, 4) = 42;
});
Q.wait();

但是，这种解决方法并不可靠。这取决于DPC++编译器没有意识到auto ptr缺少的常量（事实上，编译器拒绝ArrayWrapper<int>* ptr，但不拒绝auto ptr）。

赞(0）回复(0）举报 2023-06-07

我来回答

如何在SYCL 2020 / DPC++中实现自定义4维数组查看器/ Package 器？

3条答案

相关问题

热门标签

最新问答