缓冲区内的 SYCL 缓冲区

Posted 2023-03-10

【中文标题】缓冲区内的 SYCL 缓冲区

【英文标题】：SYCL buffers within buffers

【发布时间】：2019-07-09 18:50:56

【问题描述】：

假设我有以下包含缓冲区的结构：

struct SomeAllocatorCode 
  int* rawData;
  size_t rawDataSize;
  cl::sycl::buffer<int> nestedBuffer;
  SomeAllocatorCode(int* rawData, size_t size): rawData(rawData),
  rawDataSize(rawDataSize), nestedBuffer(rawData, cl::sycl::range<1>(rawDataSize)) 
;

然后我想创建一个像这样的缓冲区：

    int* data = new int[64];
    SomeAllocatorCode* allocator = new SomeAllocatorCode(data, 64);
    cl::sycl::buffer<SomeAllocatorCode> topLevelBuffer(allocator, 
         cl::sycl::range<1>(1));

我将如何从设备代码中读取nestedBuffer？是否可以像这样构造数据？像往常一样使用数据访问器并在使用读取访问器访问topLevelBuffer 后简单地获取nestedBuffer 的数据访问器就足够了吗？

【参考方案1】：

我建议您了解内存缓冲区和访问器在 SYCL 中的工作方式，这将帮助您采用最佳方法来完成您正在尝试做的事情。 以下是一些有用资源的链接：

Basics on buffers and accessors

Memory guide

Code sample showing good practice

【讨论】：

我一直在尝试的方法是调整结构层次结构，以便父结构保持 sycl::buffer 的视图到子结构。每个子结构都做同样的事情。结构的视图与原始结构相同，但缓冲区被访问器替换。这种方法的问题是访问器不可为空，并且在运行时复制到无效初始化的访问器会失败。有没有办法使用占位符来解决这个问题？可空访问器的最佳方法是什么？

访问器可以为空是必需的，因为要获得设备访问器，我们需要命令组处理程序。此时，缓冲区必须已经定义。要定义缓冲区，我们需要初始化内存。要初始化设备访问器的内存，我们需要命令组处理程序（除非它们可以为空）。这个循环中是否有任何点可以适应支持多级间接？看起来可以为空的访问器是最简单的，因为它们的内存可以在设备内核之前在命令组处理程序中初始化。

【参考方案2】：

这是一个如何使用嵌套占位符访问器解决此问题的示例：Possible ComputeCPP SYCL bug reading nested buffers 如该答案所述，位于设备缓冲区中的以下访问器不起作用，因此目前无法进行多级间接。

【讨论】：

这里有一个更复杂的 SYCL 用法供参考：github.com/BenjaminTrapani/FunGPU/blob/master/Core/… 它处理设备端分配，并通过将所有内容存储在固定大小的内存块中来避免多重间接问题。