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

Question

我是 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 中的指针是指向主机内存的指针，设备无权访问。

Answer 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 分配器可以传递给容器。