RuntimeError: Host accessors cannot be created for buffers which are currently in use by a command graph.

[MOLOjl]

RuntimeError: Host accessors cannot be created for buffers which are currently in use by a command graph.

I'm trying to develop a PyTorch extension for oneDNN. When I pass SYCL buffer memory storage between different C++ calls from Python, an error occurs.

For example, in the first Python call to a C++ function, I create a buffer like this:

void* test_ptr;
{
    auto buffer_ = new sycl::buffer<uint8_t, 1>(length * sizeof(float));
    test_ptr = static_cast<void*>(buffer_);

    auto src = buffer_->get_host_access();
    uint8_t* src_ptr = src.get_pointer();
    memcpy(src_ptr, src_0_data.data(), length * sizeof(float));
}

After some kernel computation, I return this test_ptr to Python. In the next Python call, I reinterpret this pointer as a buffer and try to access it in another C++ function like this:

void printtest(void* ptr) {
    auto buffer_ = static_cast<sycl::buffer<uint8_t, 1>*>(ptr);
    auto h_access = buffer_->get_host_access();
    uint8_t* src_h_ptr = h_access.get_pointer();

    std::vector<float> dst_data(20);
    std::memcpy(dst_data.data(), src_h_ptr, 20);
    printf("print test %f\n", dst_data[0]);
}

However, the above error is thrown in auto h_access = buffer_->get_host_access();.

Is there any way to release the command graph's occupation of the buffers or another way to copy the data in the returned buffer to a host array?

Any help would be greatly appreciated.

[AlexeySachkov]

Tagging @EwanC here to help answer the question

[EwanC]

Hi, thanks for the question

Is there any way to release the command graph's occupation of the buffers or another way to copy the data in the returned buffer to a host array?

A graph's occupation of the buffer can be released by destroying the graph that is using the buffer. See https://github.com/intel/llvm/blob/sycl/sycl/test-e2e/Graph/Inputs/basic_buffer.cpp for an example of using host_accessor to read back the graph output from
a buffer after the graph has finished executing and has been destroyed.

Alternatively, you can copy the data in the returned buffer to a host-array using a handler::copy() operation. That could be as a node in the graph like in this test - https://github.com/intel/llvm/blob/sycl/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host.cpp#L32,
or you could submit it after the graph has executed with a dependency on the graph execution.

[MOLOjl]

thanks for your help, but it'still couldn't work

A graph's occupation of the buffer can be released by destroying the graph that is using the buffer. See https://github.com/intel/llvm/blob/sycl/sycl/test-e2e/Graph/Inputs/basic_buffer.cpp for an example of using host_accessor to read back the graph output from a buffer after the graph has finished executing and has been destroyed.

I tried to construct a graph like you mentioned above, but get an error:
RuntimeError: ext_oneapi_hip backend is not supported by SYCL Graph extension.

The cgh(command graph handler) in sycl seems hidden for developer, most methods(even constructor and deconstructor) are private. In oneDNN, all the task was submit to sycl queue with the default cgh.
Like this:

q->submit([&](::sycl::handler &cgh) {
    auto buffer_ = static_cast<sycl::buffer<uint8_t, 1> *>(ptr);
    auto h_access = buffer_->get_host_access();  
    uint8_t *src_h_ptr = h_access.get_pointer();

    std::vector<float> dst_data(1*20);
    std::memcpy(dst_data.data(), src_h_ptr, 1*20);  

    printf("host_access got\n");
    printf("print test %f, h_ptr: %p \n", dst_data[0], src_h_ptr);
    delete &cgh;
});

The cgh in the lambda expression is construct and deconstruct implicit.
However, each python call will be bounded with a different cgh, so error occurs when I reuse the buffer that is used by the last cgh.
My question is:
Is there some way to explicit success the cgh or just destroy it safely? (I even tried to modify the source code and make the destructor of the sycl::handler public, but unfortunetly segment fault error occurs.)

[EwanC]

I tried to construct a graph like you mentioned above, but get an error:
RuntimeError: ext_oneapi_hip backend is not supported by SYCL Graph extension.

You should check the device you are using for if the aspect::ext_oneapi_graph aspect is present, this reports whether the device supports the extension. sycl-ls -v should show it. The minimum version of ROCm required to support sycl_ext_oneapi_graph is 5.5.1. This error may be because you are using ROCm version older than that.

Is there some way to explicit success the cgh or just destroy it safely? (I even tried to modify the source code and make the destructor of the sycl::handler public, but unfortunetly segment fault error occurs.)

In SYCL the handler class if for the runtime to manage rather than the user, so I don't think so.

In your example I think you could remove the buffer->get_host_access(); which is causing the exception to be thrown and replace it with a handler::copy. For example, something like:

q->submit([&](::sycl::handler &cgh) {
    auto buffer_ = static_cast<sycl::buffer<uint8_t, 1> *>(ptr);
    auto Acc = buffer_->get_access(cgh);
    std::vector<float> dst_data(1*20);
    cgh.copy(Acc, dst_data.data());
});