Parts of this lecture will focus primarily on the buffer/accessor model.

* Often when writing SYCL kernel functions there is initial data which has been allocated somewhere else in the application.

auto devicePtr = sycl::malloc_device(sizeInBytes, gpuQueue);

gpuQueue.memcpy(devicePtr, initialData, sizeInBytes).wait();
						

* When using the USM model (unless using system USM) pointers passed to kernel functions must be allocated by the SYCL runtime. * This means you have to copy the initial data to the USM memory allocation using `memcpy`.

auto sharedData = sycl::malloc_shared(sizeInBytes, gpuQueue);
						

* Alternatively if your device supports shared USM allocations you can allocate memory which is shared across the host and device. * Then there is no need to copy the data to the device.

auto buf = sycl::buffer{initialData, sycl::range{size}};
						

* When using the buffer/accessor model a `buffer` can manage already allocated memory or have the SYCL runtime allocate it. * To do this simply provide an initial pointer when constructing a `buffer`. * Note that the SYCL runtime is free to allocate memory and copy this into it, which can introduce an overhead.

auto buf = sycl::buffer{initialData, sycl::range{size},
  {sycl::property::buffer::use_host_ptr{}}};
						

* To prevent the runtime allocation memory you can provide the `property::buffer::use_host_ptr` property when constructing the `buffer`. * This instructs the SYCL runtime that it may not allocate any additional memory. * Though note that the backend (such as OpenCL) may still allocate memory.

* A `buffer` will synchronize the latest modified copy of the data it manages back to the initial pointer on destruction.

auto buf = sycl::buffer{initialData, sycl::range{size}};

buf.set_final_data(finalData);
						

* To change the destination that a `buffer` will synchronize to on destruction you can call `set_final_data` with another. * The address provided must be capable of holding the size of the data the `buffer` manages.

auto buf = sycl::buffer{initialData, sycl::range{size}};

buf.set_final_data(nullptr);
						

* Alternatively to prevent the `buffer` from synchronizing back to the initial data entirely you can call `set_final_data` with `nullptr`. * A `buffer` with no final data address is useful because the data can left on a device and not copied back to the host from the device.

auto buf = sycl::buffer<T>{sycl::range{size}};
						

* As we've seen in the USM model all memory is allocated initialized, but `buffer`s can be constructed without initial data. * A `buffer` like this is called uninitialized. * To do this simply construct a `buffer` without initial data. Just remember to explicitly specify the buffer's data type as it can't be inferred from the initial data any more. * Uninitialized `buffer`s are useful for a couple of reasons because they can be allocated directly on a device and don't require moving data from the host.

* Pinned memory is a feature supported by most SYCL backends and devices. * It allows you to allocate memory which can be mapped between the host and device more efficiently, providing similar benefits to USM. * Though the requirements can vary from one device to another. * It's always best to check the vendor's programming guide.

* The SYCL runtime will always aim to manage the memory for you in the most efficient way for the target device. * Generally there are two approaches to facilitate pinned memory: * Allocate memory according to the vendor's programming guide, usually involved allocating a size of a particular multiple and aligned to a particular size, and then use the `property::buffer::use_host_ptr` property. * Create an uninitialized `buffer` and allow the runtime to allocate the memory the appropriate way.

Code_Exercises/Advanced_Data_Flow/source

Write a SYCL application which uses uninitialized `buffer`s and disabling write back.