Fusing USM and buff acc chapters

Author: Hugh Delaney

Code_Exercises/Exercise_03_Scalar_Add/solution.cpp

@@ -19,7 +19,33 @@
 
 class scalar_add;
 
-TEST_CASE("scalar_add", "scalar_add_solution") {
+TEST_CASE("scalar_add_usm", "scalar_add_solution") {
+  int a = 18, b = 24, r = 0;
+
+  auto defaultQueue = sycl::queue{};
+
+  auto dev_A = sycl::malloc_device<int>(1, defaultQueue);
+  auto dev_B = sycl::malloc_device<int>(1, defaultQueue);
+  auto dev_R = sycl::malloc_device<int>(1, defaultQueue);
+
+  defaultQueue.memcpy(dev_A, &a, 1 * sizeof(int));
+  defaultQueue.memcpy(dev_B, &b, 1 * sizeof(int));
+
+  defaultQueue
+      .submit([&](sycl::handler &cgh) {
+        cgh.single_task([=] { dev_R[0] = dev_A[0] + dev_B[0]; });
+      });
+
+  defaultQueue.memcpy(&r, dev_R, 1 * sizeof(int)).wait();
+
+  sycl::free(dev_A, defaultQueue);
+  sycl::free(dev_B, defaultQueue);
+  sycl::free(dev_R, defaultQueue);
+
+  REQUIRE(r == 42);
+}
+
+TEST_CASE("scalar_add_buff_acc", "scalar_add_solution") {
   int a = 18, b = 24, r = 0;
 
   auto defaultQueue = sycl::queue{};
@@ -35,7 +61,7 @@ TEST_CASE("scalar_add", "scalar_add_solution") {
           auto accB = sycl::accessor{bufB, cgh, sycl::read_only};
           auto accR = sycl::accessor{bufR, cgh, sycl::write_only};
 
-          cgh.single_task<scalar_add>([=] { accR[0] = accA[0] + accB[0]; });
+          cgh.single_task([=] { accR[0] = accA[0] + accB[0]; });
         })
         .wait();
   }

Code_Exercises/Exercise_03_Scalar_Add/source.cpp

@@ -6,18 +6,65 @@
 
  You should have received a copy of the license along with this
  work.  If not, see <http://creativecommons.org/licenses/by-sa/4.0/>.
+ 
+ * SYCL Quick Reference
+ * ~~~~~~~~~~~~~~~~~~~~
+ *
+ * // Include SYCL header
+ * #include <CL/sycl.hpp>
+ *
+ * // Default construct a queue
+ * auto q = sycl::queue{};
+ *
+ * // Allocate device memory
+ * auto * devPtr = sycl::malloc_device<int>(mycount, q);
+ *
+ * // Memcpy
+ * q.memcpy(dst, src, sizeof(T)*n).wait();
+ * // (dst and src are pointers)
+ *
+ * // Free memory
+ * sycl::free(ptr, q);
+ *
+ * // Construct a buffer of size n associated with ptr
+ * auto buf = sycl::buffer{ptr, sycl::range{n}};
+ *
+ * // Submit a kernel
+ * q.submit([&](sycl::handler &cgh) {
+ *    cgh.single_task([=](){
+ *      // Some kernel code
+ *      });
+ * }).wait();
+ *    
+ * // Construct an accessor for buf 
+ * // (must be done within command group)
+ *     auto acc = sycl::accessor{buf, cgh};
+ *     auto acc = sycl::accessor{buf, cgh, sycl::read_only};
+ *     auto acc = sycl::accessor{buf, cgh, sycl::write_only};
+ *     auto acc = sycl::accessor{buf, cgh, sycl::no_init};
+ *
 */
 
 #define CATCH_CONFIG_MAIN
 #include <catch2/catch.hpp>
 
-TEST_CASE("scalar_add", "scalar_add_source") {
+TEST_CASE("scalar_add_usm", "scalar_add_source") {
 
   int a = 18, b = 24, r = 0;
 
-  // Task: Compute a+b on the SYCL device
+  // Task: Compute a+b on the SYCL device using USM
   r = a + b;
 
+  REQUIRE(r == 42);
+}
+
+TEST_CASE("scalar_add_buff_acc", "scalar_add_source") {
+
+  int a = 18, b = 24, r = 0;
+
+  // Task: Compute a+b on the SYCL device using the buffer
+  // accessor memory model
+  r = a + b;
 
   REQUIRE(r == 42);
 }