Modernize accessor creation in documentation

Author: fknorr

README.md

@@ -28,17 +28,16 @@ Celerity without much hassle. If you know SYCL already, this will probably
 look very familiar to you:
 
 ```cpp
-celerity::buffer<float, 1> buf(celerity::range<1>(1024));
+celerity::buffer<float> buf{celerity::range<1>{1024}};
 queue.submit([=](celerity::handler& cgh) {
-  auto acc = buf.get_access<sycl::access::mode::discard_write>(
-    cgh,
-    celerity::access::one_to_one()              // 1
-  );
-  cgh.parallel_for<class MyKernel>(
-    celerity::range<1>(1024),                   // 2
-    [=](celerity::item<1> item) {               // 3
-      acc[item] = sycl::sin(item[0] / 1024.f);  // 4
-    });
+    celerity::accessor acc{buf, cgh,
+        celerity::access::one_to_one{},               // 1
+        celerity::write_only, celerity::no_init};
+    cgh.parallel_for<class MyKernel>(
+        celerity::range<1>{1024},                     // 2
+        [=](celerity::item<1> item) {                 // 3
+            acc[item] = sycl::sin(item[0] / 1024.f);  // 4
+        });
 });
 ```
 

docs/host-tasks.md

@@ -24,15 +24,14 @@ cgh.host_task(celerity::on_master_node, []{ ... });
 
 Buffers can be accessed in the usual fashion, although there is no `item` structure passed into the kernel. Instead,
 when constructing an accessor, a `fixed` or `all` range mapper is used to specify the range of interest. Also,
-the access target must be set to `host_buffer` explicitly in the call to `get_access`:
+the `*_host_task` selector must be used for selecting the access mode.
 
 ```cpp
 celerity::distr_queue q;
 celerity::buffer<float, 1> result;
 q.submit([=](celerity::handler &cgh) {
-    auto acc = result.get_access<celerity::access_mode::read,
-            celerity::target::host_task>(cgh,
-            celerity::access::all());
+	celerity::accessor acc{buffer, cgh, celerity::access::all{},
+			celerity::read_only_host_task};
     cgh.host_task(celerity::on_master_node, [=]{
         printf("The result is %g\n", acc[0]);
     });
@@ -148,9 +147,9 @@ splitting them along the first (slowest) dimension into contiguous memory portio
 celerity::distr_queue q;
 celerity::buffer<float, 2> buf;
 q.submit([=](celerity::handler& cgh) {
-    auto acc = buffer.get_access<celerity::access_mode::read,
-            celerity::target::host_task>(cgh,
-            celerity::experimental::access::even_split<2>());
+	celerity::accessor acc{buffer, cgh,
+			celerity::experimental::access::even_split<2>{},
+			celerity::read_only_host_task};
     cgh.host_task(celerity::experimental::collective,
             [=](celerity::experimental::collective_partition part) {
         auto aw = acc.get_allocation_window(part);

docs/pitfalls.md

@@ -50,8 +50,8 @@ celerity::distr_queue q;
 celerity::buffer<int, 1> buf;
 bool flag = false;
 q.submit(celerity::allow_by_ref, [=, &flag](celerity::handler &cgh) {
-    auto acc = buf.get_access<celerity::access_mode::read,
-        celerity::target::host_task>(cgh, celerity::access::all<1>());
+	celerity::accessor acc{buffer, cgh, celerity::access::all{},
+			celerity::read_only_host_task};
     cgh.host_task(celerity::on_master_node, [=, &flag] {
         flag = acc[0] == 42;
     });

docs/range-mappers.md

@@ -31,24 +31,24 @@ vice-versa.
 
 ### Usage
 
-Range mappers are passed as the second argument into a call to
-`celerity::buffer::get_access`. This means that the spatial accessing
+Range mappers are passed as the third argument when constructing a
+`celerity::accessor`. This means that the spatial accessing
 behavior of a kernel can vary from buffer to buffer. For example, the
 following command group specifies two different range mappers (whose
 definition is omitted) for buffers `buf_a` and `buf_b`:
 
 ```cpp
 queue.submit([=](celerity::handler& cgh) {
-    auto r_a = buf_a.get_access<celerity::access_mode::read>(cgh, my_mapper);
-    auto dw_b = buf_b.get_access<celerity::access_mode::discard_write>(cgh, other_mapper);
+	celerity::accessor r_a{cgh, buf_a, my_mapper, celerity::read_only};
+	celerity::accessor dw_b{cgh, buf_b, other_mapper, celerity::write_only, celerity::no_init};
 
     cgh.parallel_for<>(...);
 });
 ```
 
 > **NOTE**: In Celerity 0.1, range mappers were only used for compute kernels.
 > For master-node tasks (then called master-access tasks), explicit buffer ranges
-> were passed to `get_access`. These APIs have been unified and range mappers
+> were passed to `buffer::get_access`. These APIs have been unified and range mappers
 > are now required in all cases. In master node tasks, the `all` and `fixed`
 > mappers provide equivalent functionality to explicit ranges.
 

docs/reductions.md

@@ -13,7 +13,7 @@ The following distributed program computes the sum from 0 to 999 in `sum_buf` us
 celerity::distr_queue q;
 celerity::buffer<size_t, 1> sum_buf{{1}};
 q.submit([=](celerity::handler& cgh) {
-    auto rd = celerity::reduction(sum_buf, cgh, cl::sycl::plus<size_t>{},
+    auto rd = celerity::reduction(sum_buf, cgh, sycl::plus<size_t>{},
                                   celerity::property::reduction::initialize_to_identity{});
     cgh.parallel_for(celerity::range<1>{1000}, rd,
                      [=](celerity::item<1> item, auto& sum) { sum += item.get_id(0); });

docs/tutorial.md

@@ -170,8 +170,8 @@ aforementioned buffer accessors. Let's create these now - replace the TODO
 before the kernel function with the following:
 
 ```cpp
-auto r_input = input_buf.get_access<celerity::access_mode::read>(cgh, celerity::access::neighborhood(1, 1));
-auto dw_edge = edge_buf.get_access<celerity::access_mode::discard_write>(cgh, celerity::access::one_to_one());
+celerity::accessor r_input{input_buf, cgh, celerity::access::neighborhood{1, 1}, celerity::read_only};
+celerity::accessor dw_edge{edge_buf, cgh, celerity::access::one_to_one{}, celerity::write_only, celerity::no_init};
 ```
 
 If you have worked with SYCL before, these buffer accessors will look
@@ -183,8 +183,8 @@ the previous contents of `edge_buf`, which is why we choose to discard them
 and use the `discard_write` access mode.
 
 So far everything works exactly as it would in a SYCL application. However,
-there is a second parameter passed into the `celerity::buffer::get_access`
-function that is not present in its SYCL counterpart. In fact, this parameter
+there is an additional parameter passed into the `accessor`
+constructor that is not present in its SYCL counterpart. In fact, this parameter
 represents one of Celerity's most important API additions: While access modes
 tell the runtime system how a kernel intends to access a buffer, it does not
 include any information about _where_ a kernel will access said buffer. In
@@ -252,7 +252,7 @@ your `main()` function:
 
 ```cpp
 queue.submit([=](celerity::handler& cgh) {
-    auto out = edge_buf.get_access<celerity::access_mode::read, celerity::target::host_task>(cgh, celerity::access::all());
+	celerity::accessor out{edge_buf, cgh, celerity::access::all{}, celerity::read_only_host_task};
     cgh.host_task(celerity::on_master_node, [=]() {
         stbi_write_png("result.png", img_width, img_height, 1, out.get_pointer(), 0);
     });