Explicit copy from host buffer to device buffer uses uninitialized data

SYCL development
2

Hi psalz,

Yes, that’s right, so in the recent SYCL 1.2.1 revision, a clarification has been made which explicitly permits the explicit asynchronous copy operations now permits accessors of different element types, dimensionalities, access modes and access targets.

However, the restriction that host accessors must be used outside of a command group and device accessors must be used inside of a command group remains, as host accessors provide immediate blocking access for the host application on construction and releases it on destruction. Perhaps we should make the specification clearer on this.

So the issue in the code example above is that the source accessor in the copy operation is a host accessor which is not permitted so the resulting behaviour is undefined.

It’s also best not to think about buffers in terms of host buffers and device buffers. Buffers in SYCL represent data that can exist on the host and/or on any number of devices, and ther requirements for the data to be available for a given access mode in any given place is specified using accessors. This abstraction is designed to allow implementations the freedom to move data around (or not) to most efficiently meet the requirements of each command group.

For your use case, if you are simply looking to take some data initialized on the host and copy this to a buffer being accessed on the device then a way to achieve this is to use the copy operation which takes a host pointer as the source parameter and the device accessor as the destination parameter.

cl::sycl::buffer<size_t, 1> device_buf(range);

queue.submit([&](cl::sycl::handler& cgh) {
  auto dw_a = device_buf.get_access<cl::sycl::access::mode::discard_write>(cgh);
  cgh.copy(host_data.data(), dw_a);
});

However, if you need already have two buffers and you need to copy data from one to the other you should use the copy operation you have there but with two device accessors, and the SYCL runtime will copy from one to the other regardless of where the latest copy of the data resides.

cl::sycl::buffer<size_t, 1> src_buf(host_data.data(), range);
cl::sycl::buffer<size_t, 1> dest_buf(range);

queue.submit([&](cl::sycl::handler& cgh) {
  auto src_a = src_buf.get_access<cl::sycl::access::mode::read>(cgh);
  auto dest_a = dest_buf.get_access<cl::sycl::access::mode::discard_write>(cgh);
  cgh.copy(src_a , dest_a );
});

I hope this helps,

Gordon