[SYCL] Test reduction for complex numbers.

SYCL/Reduction/reduction_complex_nums.cpp

@@ -0,0 +1,58 @@
+// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple %s -o %t.out
+// RUN: %CPU_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+// RUN: %ACC_RUN_PLACEHOLDER %t.out
+
+#include <algorithm>
+#include <complex>
+#include <numeric>
+
+#include <sycl/sycl.hpp>
+
+using namespace sycl;
+
+#define BUFFER_SIZE 255
+
+// Currently, Identityless reduction for complex numbers is
+// only valid for plus operator.
+// TODO: Extend this test case once we support known_identity for std::complex
+// and more operators (apart from plus).
+template <typename T>
+void test_identityless_reduction_for_complex_nums(queue &q) {
+  // Allocate and initialize buffer on the host with all 1's.
+  buffer<std::complex<T>> valuesBuf{BUFFER_SIZE};
+  {
+    host_accessor a{valuesBuf};
+    T n = 0;
+    std::generate(a.begin(), a.end(), [&n] {
+      n++;
+      return std::complex<T>(n, n + 1);
+    });
+  }
+
+  // Buffer to hold the reduction results.
+  std::complex<T> sumResult = 0;
+  buffer<std::complex<T>> sumBuf{&sumResult, 1};
+
+  q.submit([&](handler &cgh) {
+    accessor inputVals{valuesBuf, cgh, sycl::read_only};
+    auto sumReduction = reduction(sumBuf, cgh, plus<std::complex<T>>());
+
+    cgh.parallel_for(nd_range<1>{BUFFER_SIZE, BUFFER_SIZE}, sumReduction,
+                     [=](nd_item<1> idx, auto &sum) {
+                       sum += inputVals[idx.get_global_id(0)];
+                     });
+  });
+
+  assert(sumBuf.get_host_access()[0] == std::complex<T>(32640, 32895));
+}
+
+int main() {
+  queue q;
+
+  test_identityless_reduction_for_complex_nums<float>(q);
+  if (q.get_device().has(aspect::fp64))
+    test_identityless_reduction_for_complex_nums<double>(q);
+
+  return 0;
+}