Passing sycl buffers as arguments

[eddie_patton]

I'd like to pass sycl buffers as function arguments. 

The function below doesn't work because it does not initialize the sycl buffer passed to it. 

void initUint8SyclBuffer1(queue &q,
                      buffer<uint8_t> &u8_buffer, // input and output
                      int sz, uint8_t value)
{
  range numItems{sz};

  try
  {  
      //q.submit([&u8_buffer, width, height, value](
      //          sycl::handler& h) {
      q.submit([&](auto &h) {
      accessor u8(u8_buffer, h, write_only, no_init);
      h.parallel_for(numItems, [=](auto idx) {
                        u8[idx] = value;
                    });
      });
  } catch (std::exception const &e) {
    cout << "initUint8SyclBuffer exception: " << e.what() << std::endl;
    terminate();
  }
}

whereas if I pass a vector and declare the buffer inside the function, the output is correct.

void initUint8SyclBuffer(queue &q,
                      vector<uint8_t> &u8_vector, // input and output
                      int sz, uint8_t value)
{
  range numItems{sz};

  buffer u8_buffer(u8_vector);
  try
  {  
      q.submit([&](auto &h) {
        accessor u8(u8_buffer, h, write_only, no_init);
        h.parallel_for(numItems, [=](auto idx) {
                          u8[idx] = value;
                      });
      });
  } catch (std::exception const &e) {
    cout << "initUint8SyclBuffer exception: " << e.what() << std::endl;
    terminate();
  }
}

The above functions are called from main

  int sz = 10; 
  uint8_t initValue = 128;
  std::vector<uint8_t> u8_data(sz);

  //buffer<uint8_t, 1> u8_data_buffer{u8_data.data(), width * height};
  buffer u8_data_buffer(u8_data);
  queue sycl_que(selector, exception_handler);
  try {
    // Print out the device information used for the kernel code.
    cout << "Running on device: "
         << sycl_que.get_device().get_info<info::device::name>() << "\n";
    
    //initUint8SyclBuffer(sycl_que, u8_data, sz, initValue);    // This works
    initUint8SyclBuffer1(sycl_que, u8_data_buffer, sz, initValue);  // This does not

  } catch (std::exception const &e) {
    cout << "An exception is caught on device:  " << e.what() << std::endl;
    terminate();
  }

  sycl_que.wait();

I've attached a zip of the entire project I'm running using visual studio code and compiling using the vscode powershell. It contains a readme on how it's compiled and run. 

[cw_intel · Accepted solution]

When using a SYCL buffer, the ownership of the pointer passed to the constructor of the class is, by default, passed to SYCL runtime, and that pointer cannot be used on the host side until the buffer is destroyed. A SYCL application can access the contents of the memory managed by a SYCL buffer by using a host_accessor.  For the details, please refer to the SYCL 2020 spec,  in section 4.7.4. 

So, if a buffer creation happens within a separate function scope, when execution advances beyond this function scope, buffer destructor is invoked which relinquishes the ownership of data and copies back the data to the host memory, and you can access the vector memory directly. 

For your code, where you pass a sycl buffer to a function,  you can use host_accessor to access the data. for example,

  host_accessor h(u8_data_buffer,read_only);
  for(int i = 0; i < sz; i++) { 
      if(h[i] != initValue) {
...
      }
    }

Thanks.

[eddie_patton]

Thanks CW. That really helps. 
Your comment also made me realize that all I had to do was add brackets to force u8_buffer_data to be out of scope when u8_data is used (and thus force a sync between host mem and device mem). 

int main() {

  // Create device selector for the device of your interest.

  // The default device selector will select the most performant device.

  auto selector = default_selector_v;

  cout << "Starting main" << std::endl;

  int sz = 10;

  uint8_t initValue = 128;

  std::vector<uint8_t> u8_data(sz);

  {

    // Create sycl u8_data_buffer in this scope in order to make get the host memeory to

    // sync with device memory when the u8_data_buffer goes out of scope.

    queue sycl_que(selector, exception_handler);

    buffer u8_data_buffer(u8_data);

    //buffer<uint8_t, 1> u8_data_buffer{u8_data.data(), range<1>(sz)};

    try {

      // Print out the device information used for the kernel code.

      cout << "Running on device: "

          << sycl_que.get_device().get_info<info::device::name>() << "\n";

      //initUint8SyclBuffer(sycl_que, u8_data, sz, initValue);    // This works

      initUint8SyclBuffer1(sycl_que, u8_data_buffer, sz, initValue);  // This does not

    } catch (std::exception const &e) {

      cout << "An exception is caught on device:  " << e.what() << std::endl;

      terminate();

    }

    sycl_que.wait();

  }  // u8_data_buffer device mem will now sync to host mem

  //host_accessor h_u8_data(u8_data_buffer,read_only);

  for(int i = 0; i < sz; i++) {

      cout << (int)u8_data[i] << ", ";

      //if(h_u8_data[i] != initValue) {

      if(u8_data[i] != initValue) {

        cerr << "ERROR: u8_data[" << i << "] = " << (int)u8_data[i] <<

               " does not equal init value " << (int)initValue << "." << std::endl;

        exit(1);

      }

  }

  cout << std::endl;