为自定义对象 oneAPI 创建数据缓冲区时遇到问题

Posted 2023-02-22

【中文标题】为自定义对象 oneAPI 创建数据缓冲区时遇到问题

【英文标题】：Having trouble creating data buffers for custom objects oneAPI

【发布时间】：2020-05-08 14:56:01

【问题描述】：

我是 oneAPI 和类似框架的新手，因此在使用 SYCL 数据缓冲区进行数据管理时遇到了麻烦。

我的任务是使用 Aho-Corasick 算法在给定字符串中查找子字符串。

我的想法是构建一个 trie，然后提交一个内核，该内核将在 trie 中并行查找子字符串。因此，为此我创建了一个 SYCL 队列，为字符串（用于查找子字符串的那个）、向量（用于存储搜索结果）和我的 Aho-Corasick 对象创建了缓冲区，该对象包含先前构建的特里树的根.但是，关于最后一个我不确定，因为我正在为主机内存中的一个对象创建一个缓冲区，该缓冲区包含指向其他对象的指针（例如节点，它包含指向其他节点的指针）。

Node对象的结构：

class Node 

        typedef Node *node_ptr;

    private:

        std::set<std::pair<int, std::string>> retVals;
        std::unordered_map<char, node_ptr> children;
        node_ptr fail;
        char value;

这是搜索方法：

 void
        matchWords(char *text, int startIdx, int endIdx,  cl::sycl::cl_int *matched) 

            node_ptr child = start;
            int item = startIdx;
            for (int i = startIdx; i < endIdx; ++i) 
                child = child->nextNode(text[i]);
                if (child == nullptr) 
                    child = start;
                    continue;
                
                for (const auto &returns: child->getRetVals()) 
                    matched[item++] = returns.first;
                    if (item == endIdx) item = startIdx;
                
            
        

缓冲区：

cl::sycl::buffer<char, 1> fasta_buf(tempFasta.data(), cl::sycl::range<1>(len));
cl::sycl::buffer<cl::sycl::cl_int, 1> vec_buf(vec.data(), cl::sycl::range<1>(len));
cl::sycl::buffer<aho_corasick::AhoCorasick, 1> aho_buf(a, cl::sycl::range<1>(1));

并排队求和：

q.submit([&](cl::sycl::handler &cgh) 
        auto string_acc = fasta_buf.get_access<cl::sycl::access::mode::read>(cgh);
        auto vec_acc = vec_buf.get_access<cl::sycl::access::mode::read_write>(cgh);
        auto aho_acc = aho_buf.get_access<cl::sycl::access::mode::read>(cgh);

        cgh.parallel_for<class dummy>(
                cl::sycl::range<1>(10), [=](cl::sycl::item<1> i) 
                    // 10 is the number of workers I want 
                    int startInx = (int) (i.get_linear_id() * (len / 10)); 
                    int endInx = (int) ((i.get_linear_id() + 1) * (len / 10));
                    aho_acc.get_pointer()->matchWords(string_acc.get_pointer(), startInx, endInx, vec_acc.get_pointer());
                );
    );
    q.wait_and_throw();

我发现程序在尝试访问子地图的项目后失败了。因此，我认为问题在于存储在 map 中的指针是指向主机内存的指针，设备无权访问。

【问题讨论】：

嗨@Nazar，我们将提交一份工单让工程部门查看此问题。但是，现在是周末，所以我们可能要到下周初才有人看。

另外，您使用的是 Beta05 还是 Beta06？ software.intel.com/content/www/us/en/develop/tools/…

感谢您的回复@ChileAddict-Intel，我似乎无法找到有关我使用的是 Beta05 还是 Beta06 的信息。不过，我在 3 月左右安装了 oneAPI Base Toolkit。

您将运行 Beta05（或更早版本），因为 Beta06 刚刚发布。你可以去我上面的帖子。此外，如果您想早点看到，这里是 intel.com 上的 dpc++ 论坛。 software.intel.com/en-us/forums/oneapi-data-parallel-c-compiler - 在下载页面上，您要选择英特尔 oneAPI 工具包链接 - DPC++ 是其中的一部分。

【参考方案1】：

如果我理解正确，您正在尝试在设备代码中使用std::unordered_map、std::string 和std::set。我不是英特尔专用 oneAPI SYCL 扩展方面的专家，但在纯 SYCL 1.2.1 中这是不允许的，如果这在 DPC++ 中有效，我会感到惊讶。

SYCL 1.2.1 规范并未真正定义 SYCL 如何与标准库交互。虽然某些实现可能能够对标准库的某些定义良好的部分在设备代码中作为扩展工作（通常例如std::数学函数）做出一些保证，但这并不能在 SYCL 实现中得到普遍保证。 另外在设备代码中支持 STL 容器（SYCL 规范不要求）我认为这特别困难，而且我从未听说过支持它的 SYCL 实现。这是因为容器通常使用 SYCL 设备代码中不支持的机制，因为它们需要运行时支持，例如抛出异常。因为在 GPU 上没有 C++ 运行时，所以这样的机制不能在 SYCL 中工作。

同样重要的是要理解，这并不是真正的 SYCL 特定限制，而是异构编程模型中的常见限制。其他异构编程模型（例如 CUDA）出于类似原因施加了类似的限制。

内核中容器的另一个困难是，STL 数据结构通常不是真正为 SYCL 设备上的大规模并行 SIMT 执行模型设计的，这使得它们容易出现竞争条件。

最后一个问题是您已经确定的问题：您正在将指针复制到主机内存。由于您使用的是 oneAPI DPC++，因此使用基于指针的数据结构的最简单解决方案是使用英特尔 SYCL 扩展 unified shared memory (USM)，它可用于生成在主机和设备上均有效的指针。如果设备代码支持，还有一个 USM 分配器可以传递给容器。

【讨论】：

我同意这个答案。如果你想避免使用指针，你可以定义一些看起来像指针但在场景后面在缓冲区中偏移的接口。参见例如github.com/keryell/ronan/raw/gh-pages/Talks/2016/…