From 243d8c1a0113ea37de44e47a5ede72c5a18ca615 Mon Sep 17 00:00:00 2001
From: Jefferson Le Quellec <jefferson.lequellec@codeplay.com>
Date: Tue, 14 Mar 2023 17:05:41 +0100
Subject: [PATCH 1/6] Add helper functions for marray complex testing

---
 SYCL/Complex/sycl_complex_helper.hpp | 36 ++++++++++++++++++++++++++--
 1 file changed, 34 insertions(+), 2 deletions(-)

diff --git a/SYCL/Complex/sycl_complex_helper.hpp b/SYCL/Complex/sycl_complex_helper.hpp
index de4cc1813a..4b018fcd0d 100644
--- a/SYCL/Complex/sycl_complex_helper.hpp
+++ b/SYCL/Complex/sycl_complex_helper.hpp
@@ -33,7 +33,8 @@ template <> const char *get_typename<double>() { return "double"; }
 template <> const char *get_typename<float>() { return "float"; }
 template <> const char *get_typename<sycl::half>() { return "sycl::half"; }
 
-// Helper to test each complex specilization
+/// Helper to test each complex specilization
+
 // Overload for cplx_test_cases
 template <template <typename> typename action, typename... argsT>
 bool test_valid_types(sycl::queue &Q, argsT... args) {
@@ -195,8 +196,17 @@ template <> auto constexpr init_deci(sycl::half re) {
   return static_cast<float>(re);
 }
 
-// Helper for comparing SyclCPLX and standard c++ results
+template <typename Tout, typename Tin, std::size_t NumElements>
+auto constexpr convert_marray(sycl::marray<std::complex<Tin>, NumElements> c) {
+  sycl::marray<std::complex<Tout>, NumElements> rtn;
+  for (std::size_t i = 0; i < NumElements; ++i)
+    rtn[i] = c[i];
+  return rtn;
+}
+
+/// Helper for comparing SyclCPLX and standard c++ results
 
+// Complex overloads
 template <typename T>
 bool check_results(experimental::complex<T> output, std::complex<T> reference,
                    bool is_device) {
@@ -210,7 +220,19 @@ bool check_results(experimental::complex<T> output, std::complex<T> reference,
   }
   return true;
 }
+template <typename T, std::size_t NumElements>
+bool check_results(sycl::marray<experimental::complex<T>, NumElements> output,
+                   sycl::marray<std::complex<T>, NumElements> reference,
+                   bool is_device) {
+  for (std::size_t i = 0; i < NumElements; ++i) {
+    if (!check_results(output[i], reference[i], is_device)) {
+      return false;
+    }
+  }
+  return true;
+}
 
+// Scalar overloads
 template <typename T>
 bool check_results(T output, T reference, bool is_device) {
   if (!almost_equal_scalar(output, reference, SYCL_CPLX_TOL_ULP)) {
@@ -223,3 +245,13 @@ bool check_results(T output, T reference, bool is_device) {
   }
   return true;
 }
+template <typename T, std::size_t NumElements>
+bool check_results(sycl::marray<T, NumElements> output,
+                   sycl::marray<T, NumElements> reference, bool is_device) {
+  for (std::size_t i = 0; i < NumElements; ++i) {
+    if (!check_results(output[i], reference[i], is_device)) {
+      return false;
+    }
+  }
+  return true;
+}

From f0ac5e66fe9759814de7c25a6f01b43e4636beb8 Mon Sep 17 00:00:00 2001
From: Jefferson Le Quellec <jefferson.lequellec@codeplay.com>
Date: Tue, 14 Mar 2023 17:06:17 +0100
Subject: [PATCH 2/6] Fix warning messages

---
 SYCL/Complex/sycl_complex_math_test_cases.hpp | 4 ++--
 1 file changed, 2 insertions(+), 2 deletions(-)

diff --git a/SYCL/Complex/sycl_complex_math_test_cases.hpp b/SYCL/Complex/sycl_complex_math_test_cases.hpp
index 4cfd29a2e9..2318f51950 100644
--- a/SYCL/Complex/sycl_complex_math_test_cases.hpp
+++ b/SYCL/Complex/sycl_complex_math_test_cases.hpp
@@ -58,8 +58,8 @@ template <template <typename> typename test_struct> struct cplx_test_cases {
     }
 
     for (auto &test_tuple : comp_test_values) {
-      test_passes &= test_valid_types<test_struct>(Q, get<0>(test_tuple),
-                                                   get<1>(test_tuple),
+      test_passes &= test_valid_types<test_struct>(Q, std::get<0>(test_tuple),
+                                                   std::get<1>(test_tuple),
                                                    /*use_ref*/ true);
     }
 

From 5fc59f88b81c46897278352da1d928f962f9f4c2 Mon Sep 17 00:00:00 2001
From: Jefferson Le Quellec <jefferson.lequellec@codeplay.com>
Date: Tue, 14 Mar 2023 17:07:03 +0100
Subject: [PATCH 3/6] Add test cases for marray complex

---
 .../sycl_complex_marray_test_cases.hpp        | 321 ++++++++++++++++++
 1 file changed, 321 insertions(+)
 create mode 100644 SYCL/Complex/sycl_complex_marray_test_cases.hpp

diff --git a/SYCL/Complex/sycl_complex_marray_test_cases.hpp b/SYCL/Complex/sycl_complex_marray_test_cases.hpp
new file mode 100644
index 0000000000..43436d024b
--- /dev/null
+++ b/SYCL/Complex/sycl_complex_marray_test_cases.hpp
@@ -0,0 +1,321 @@
+#include <iostream>
+#include <tuple>
+#include <vector>
+
+#include "sycl_complex_helper.hpp"
+
+#define DEFAULT_TEST_CASE_SIZE 47
+#define POLAR_TEST_CASE_SIZE 9
+#define GETTERS_TEST_CASE_SIZE 8
+
+////////////////////////////////////////////////////////////////////////////////
+/// FORWARD DECLARATION OF TESTS
+////////////////////////////////////////////////////////////////////////////////
+
+// Maths
+template <typename T> struct test_abs;
+template <typename T> struct test_arg;
+template <typename T> struct test_norm;
+
+template <typename T> struct test_acos;
+template <typename T> struct test_asin;
+template <typename T> struct test_atan;
+template <typename T> struct test_acosh;
+template <typename T> struct test_asinh;
+template <typename T> struct test_atanh;
+template <typename T> struct test_conj;
+template <typename T> struct test_cos;
+template <typename T> struct test_cosh;
+template <typename T> struct test_exp;
+template <typename T> struct test_log;
+template <typename T> struct test_log10;
+template <typename T> struct test_proj;
+template <typename T> struct test_sin;
+template <typename T> struct test_sinh;
+template <typename T> struct test_sqrt;
+template <typename T> struct test_tan;
+template <typename T> struct test_tanh;
+
+template <typename T> struct test_polar;
+
+template <typename T> struct test_pow_cplx_cplx;
+template <typename T> struct test_pow_cplx_deci;
+template <typename T> struct test_pow_deci_cplx;
+
+// Operators
+template <typename T> struct test_add;
+template <typename T> struct test_sub;
+template <typename T> struct test_mul;
+template <typename T> struct test_div;
+
+template <typename T> struct test_assign_add;
+template <typename T> struct test_assign_sub;
+template <typename T> struct test_assign_mul;
+template <typename T> struct test_assign_div;
+
+template <typename T> struct test_unary_add;
+template <typename T> struct test_unary_sub;
+
+// Getters
+template <typename T> struct test_real;
+template <typename T> struct test_imag;
+
+////////////////////////////////////////////////////////////////////////////////
+/// TEST CASES COMPLEX
+////////////////////////////////////////////////////////////////////////////////
+
+// Driver
+template <template <typename> typename test_struct>
+struct marray_cplx_test_cases {
+  static std::vector<cmplx<double>> std_test_values;
+  static std::tuple<std::vector<cmplx<double>>, std::vector<cmplx<double>>>
+      comp_test_values;
+
+  static const char *test_name;
+
+  bool operator()(sycl::queue &Q) {
+    bool test_passes = true;
+
+    if (!std_test_values.empty()) {
+      test_passes &= test_valid_types<test_struct>(Q, std_test_values);
+    }
+
+    const std::vector<cmplx<double>> init = std::get<0>(comp_test_values);
+    const std::vector<cmplx<double>> ref = std::get<1>(comp_test_values);
+
+    if (!init.empty() && !ref.empty()) {
+      test_passes &=
+          test_valid_types<test_struct>(Q, init, ref, /*use_ref*/ true);
+    }
+
+    if (!test_passes)
+      std::cerr << test_name << " failed\n";
+
+    return test_passes;
+  }
+};
+
+// Default values
+template <template <typename> typename test_struct>
+std::vector<cmplx<double>>
+    marray_cplx_test_cases<test_struct>::std_test_values = {
+        cmplx<double>(0, 1),           cmplx<double>(0, -1),
+        cmplx<double>(0, 0.5),         cmplx<double>(0, -0.5),
+        cmplx<double>(0, 0),           cmplx<double>(0, INFINITYd),
+        cmplx<double>(0, NANd),
+
+        cmplx<double>(1, 1),           cmplx<double>(1, -1),
+        cmplx<double>(1, 0.5),         cmplx<double>(1, -0.5),
+        cmplx<double>(1, 0),           cmplx<double>(1, INFINITYd),
+        cmplx<double>(1, NANd),
+
+        cmplx<double>(-1, 1),          cmplx<double>(-1, -1),
+        cmplx<double>(-1, 0.5),        cmplx<double>(-1, -0.5),
+        cmplx<double>(-1, 0),          cmplx<double>(-1, INFINITYd),
+        cmplx<double>(-1, NANd),
+
+        cmplx<double>(0.5, 1),         cmplx<double>(0.5, -1),
+        cmplx<double>(0.5, 0.5),       cmplx<double>(0.5, -0.5),
+        cmplx<double>(0.5, 0),         cmplx<double>(0.5, INFINITYd),
+        cmplx<double>(0.5, NANd),
+
+        cmplx<double>(-0.5, 1),        cmplx<double>(-0.5, -1),
+        cmplx<double>(-0.5, 0.5),      cmplx<double>(-0.5, -0.5),
+        cmplx<double>(-0.5, 0),        cmplx<double>(-0.5, INFINITYd),
+        cmplx<double>(-0.5, NANd),
+
+        cmplx<double>(INFINITYd, 1),   cmplx<double>(INFINITYd, -1),
+        cmplx<double>(INFINITYd, 0.5), cmplx<double>(INFINITYd, -0.5),
+        cmplx<double>(INFINITYd, 0),   cmplx<double>(INFINITYd, INFINITYd),
+
+        cmplx<double>(NANd, 1),        cmplx<double>(NANd, -1),
+        cmplx<double>(NANd, 0.5),      cmplx<double>(NANd, -0.5),
+        cmplx<double>(NANd, 0),        cmplx<double>(NANd, NANd),
+};
+template <template <typename> typename test_struct>
+std::tuple<std::vector<cmplx<double>>, std::vector<cmplx<double>>>
+    marray_cplx_test_cases<test_struct>::comp_test_values({}, {});
+
+/// Maths
+
+// test_abs
+template <>
+const char *marray_cplx_test_cases<test_abs>::test_name = "abs test";
+// test_arg
+template <>
+const char *marray_cplx_test_cases<test_arg>::test_name = "arg test";
+// test_norm
+template <>
+const char *marray_cplx_test_cases<test_norm>::test_name = "norm test";
+
+// test_acos
+template <>
+const char *marray_cplx_test_cases<test_acos>::test_name = "acos test";
+// test_asin
+template <>
+const char *marray_cplx_test_cases<test_asin>::test_name = "asin test";
+// test_atan
+template <>
+const char *marray_cplx_test_cases<test_atan>::test_name = "atan test";
+// test_acosh
+template <>
+const char *marray_cplx_test_cases<test_acosh>::test_name = "acos test";
+// test_asinh
+template <>
+const char *marray_cplx_test_cases<test_asinh>::test_name = "asinh test";
+// test_atanh
+template <>
+const char *marray_cplx_test_cases<test_atanh>::test_name = "atanh test";
+// test_conj
+template <>
+const char *marray_cplx_test_cases<test_conj>::test_name = "conj test";
+// test_cos
+template <>
+const char *marray_cplx_test_cases<test_cos>::test_name = "cos test";
+// test_cosh
+template <>
+const char *marray_cplx_test_cases<test_cosh>::test_name = "cosh test";
+// test_exp
+template <>
+const char *marray_cplx_test_cases<test_exp>::test_name = "exp test";
+// test_log
+template <>
+const char *marray_cplx_test_cases<test_log>::test_name = "log test";
+// test_log10
+template <>
+const char *marray_cplx_test_cases<test_log10>::test_name = "log10 test";
+// test_proj
+template <>
+const char *marray_cplx_test_cases<test_proj>::test_name = "proj test";
+// test_sin
+template <>
+const char *marray_cplx_test_cases<test_sin>::test_name = "sin test";
+// test_sinh
+template <>
+const char *marray_cplx_test_cases<test_sinh>::test_name = "sinh test";
+// test_sqrt
+template <>
+const char *marray_cplx_test_cases<test_sqrt>::test_name = "sqrt test";
+// test_tan
+template <>
+const char *marray_cplx_test_cases<test_tan>::test_name = "tan test";
+// test_tanh
+template <>
+const char *marray_cplx_test_cases<test_tanh>::test_name = "tanh test";
+
+// test_polar
+template <>
+std::vector<cmplx<double>> marray_cplx_test_cases<test_polar>::std_test_values =
+    {
+        cmplx<double>(0, 1),   cmplx<double>(0, 0.5),   cmplx<double>(0, 0),
+
+        cmplx<double>(1, 1),   cmplx<double>(1, 0.5),   cmplx<double>(1, 0),
+
+        cmplx<double>(0.5, 1), cmplx<double>(0.5, 0.5), cmplx<double>(0.5, 0),
+};
+template <>
+const char *marray_cplx_test_cases<test_polar>::test_name = "polar test";
+
+// test_pow_cplx_cplx
+template <>
+const char *marray_cplx_test_cases<test_pow_cplx_cplx>::test_name =
+    "pow cplx cplx test";
+// test_pow_cplx_deci
+template <>
+const char *marray_cplx_test_cases<test_pow_cplx_deci>::test_name =
+    "pow cplx deci test";
+// test_pow_deci_cplx
+template <>
+const char *marray_cplx_test_cases<test_pow_deci_cplx>::test_name =
+    "pow deci cplx test";
+
+/// Operators
+
+// test_add
+template <>
+const char *marray_cplx_test_cases<test_add>::test_name = "add test";
+// test_sub
+template <>
+const char *marray_cplx_test_cases<test_sub>::test_name = "sub test";
+// test_mul
+template <>
+const char *marray_cplx_test_cases<test_mul>::test_name = "mul test";
+// test_div
+template <>
+const char *marray_cplx_test_cases<test_div>::test_name = "div test";
+
+// test_assign_add
+template <>
+const char *marray_cplx_test_cases<test_assign_add>::test_name =
+    "assign add test";
+// test_assign_sub
+template <>
+const char *marray_cplx_test_cases<test_assign_sub>::test_name =
+    "assign sub test";
+// test_assign_mul
+template <>
+const char *marray_cplx_test_cases<test_assign_mul>::test_name =
+    "assign mul test";
+// test_assign_div
+template <>
+const char *marray_cplx_test_cases<test_assign_div>::test_name =
+    "assign div test";
+
+// test_unary_add
+template <>
+const char *marray_cplx_test_cases<test_unary_add>::test_name =
+    "unary add test";
+// test_unary_sub
+template <>
+const char *marray_cplx_test_cases<test_unary_sub>::test_name =
+    "unary sub test";
+
+////////////////////////////////////////////////////////////////////////////////
+/// TEST CASES SCALAR
+////////////////////////////////////////////////////////////////////////////////
+
+// Driver
+template <template <typename> typename test_struct>
+struct marray_scalar_test_cases {
+  static std::vector<double> std_test_values;
+  static std::tuple<std::vector<double>, std::vector<double>> comp_test_values;
+
+  static const char *test_name;
+
+  bool operator()(sycl::queue &Q) {
+    bool test_passes = true;
+
+    if (!std_test_values.empty()) {
+      test_passes &= test_valid_types<test_struct>(Q, std_test_values);
+    }
+
+    const std::vector<double> init = std::get<0>(comp_test_values);
+    const std::vector<double> ref = std::get<1>(comp_test_values);
+
+    if (!init.empty() && !ref.empty()) {
+      test_passes &=
+          test_valid_types<test_struct>(Q, init, ref, /*use_ref*/ true);
+    }
+
+    if (!test_passes)
+      std::cerr << test_name << " failed\n";
+
+    return test_passes;
+  }
+};
+
+// Default values
+template <template <typename> typename test_struct>
+std::vector<double> marray_scalar_test_cases<test_struct>::std_test_values = {
+    0, 0.5, 1, -0, -0.5, -1, INFINITYd, NANd,
+};
+template <template <typename> typename test_struct>
+std::tuple<std::vector<double>, std::vector<double>>
+    marray_scalar_test_cases<test_struct>::comp_test_values({}, {});
+
+// test_real
+template <>
+const char *marray_scalar_test_cases<test_real>::test_name = "real test";
+// test_imag
+template <>
+const char *marray_scalar_test_cases<test_imag>::test_name = "imag test";

From 88332d1e0d39cf993c0f549f87615cc26740171c Mon Sep 17 00:00:00 2001
From: Jefferson Le Quellec <jefferson.lequellec@codeplay.com>
Date: Tue, 14 Mar 2023 17:07:32 +0100
Subject: [PATCH 4/6] Add marray complex getter tests

---
 .../sycl_complex_marray_getters_test.cpp      | 94 +++++++++++++++++++
 1 file changed, 94 insertions(+)
 create mode 100644 SYCL/Complex/sycl_complex_marray_getters_test.cpp

diff --git a/SYCL/Complex/sycl_complex_marray_getters_test.cpp b/SYCL/Complex/sycl_complex_marray_getters_test.cpp
new file mode 100644
index 0000000000..158f0e4e9b
--- /dev/null
+++ b/SYCL/Complex/sycl_complex_marray_getters_test.cpp
@@ -0,0 +1,94 @@
+// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -fsycl-device-code-split=per_kernel %s -o %t.out
+// RUN: %CPU_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+// RUN: %ACC_RUN_PLACEHOLDER %t.out
+
+#include "sycl_complex_marray_test_cases.hpp"
+
+template <typename T> struct test_real {
+  bool operator()(sycl::queue &Q, const std::vector<double> &init,
+                  const std::vector<double> &ref = {}, bool use_ref = false) {
+    bool pass = true;
+
+    /* std::complex test cases */
+    sycl::marray<T, GETTERS_TEST_CASE_SIZE> std_in;
+    /* sycl::complex test cases */
+    sycl::marray<experimental::complex<T>, GETTERS_TEST_CASE_SIZE> cplx_in;
+
+    for (std::size_t i = 0; i < GETTERS_TEST_CASE_SIZE; ++i) {
+      std_in[i] = static_cast<T>(init[i]);
+      cplx_in[i] =
+          experimental::complex<T>{static_cast<T>(init[i]), static_cast<T>(0)};
+    }
+
+    auto *cplx_out =
+        sycl::malloc_shared<sycl::marray<T, GETTERS_TEST_CASE_SIZE>>(1, Q);
+
+    /* Check cplx::complex output from device */
+    Q.single_task([=]() { *cplx_out = cplx_in.real(); }).wait();
+    pass &= check_results(*cplx_out, std_in, /*is_device*/ true);
+
+    /* Check cplx::complex output from host */
+    *cplx_out = cplx_in.real();
+    pass &= check_results(*cplx_out, std_in, /*is_device*/ false);
+
+    sycl::free(cplx_out, Q);
+
+    return pass;
+  }
+};
+
+template <typename T> struct test_imag {
+  bool operator()(sycl::queue &Q, const std::vector<double> &init,
+                  const std::vector<double> &ref = {}, bool use_ref = false) {
+    bool pass = true;
+
+    /* std::complex test cases */
+    sycl::marray<T, GETTERS_TEST_CASE_SIZE> std_in;
+    /* sycl::complex test cases */
+    sycl::marray<experimental::complex<T>, GETTERS_TEST_CASE_SIZE> cplx_in;
+
+    for (std::size_t i = 0; i < GETTERS_TEST_CASE_SIZE; ++i) {
+      std_in[i] = static_cast<T>(init[i]);
+      cplx_in[i] =
+          experimental::complex<T>{static_cast<T>(0), static_cast<T>(init[i])};
+    }
+
+    auto *cplx_out =
+        sycl::malloc_shared<sycl::marray<T, GETTERS_TEST_CASE_SIZE>>(1, Q);
+
+    /* Check cplx::complex output from device */
+    Q.single_task([=]() { *cplx_out = cplx_in.imag(); }).wait();
+    pass &= check_results(*cplx_out, std_in, /*is_device*/ true);
+
+    /* Check cplx::complex output from host */
+    *cplx_out = cplx_in.imag();
+    pass &= check_results(*cplx_out, std_in, /*is_device*/ false);
+
+    sycl::free(cplx_out, Q);
+
+    return pass;
+  }
+};
+
+int main() {
+  sycl::queue Q;
+
+  bool test_passes = true;
+
+  /* Test real getter */
+
+  {
+    marray_scalar_test_cases<test_real> test;
+    test_passes &= test(Q);
+  }
+
+  /* Test imag getter */
+
+  {
+    marray_scalar_test_cases<test_imag> test;
+    test_passes &= test(Q);
+  }
+
+  return !test_passes;
+}

From 0a050abdab1e1109c8210663ac38352d874eddf0 Mon Sep 17 00:00:00 2001
From: Jefferson Le Quellec <jefferson.lequellec@codeplay.com>
Date: Tue, 14 Mar 2023 17:08:02 +0100
Subject: [PATCH 5/6] Add marray complex math tests

---
 .../Complex/sycl_complex_marray_math_test.cpp | 486 ++++++++++++++++++
 1 file changed, 486 insertions(+)
 create mode 100644 SYCL/Complex/sycl_complex_marray_math_test.cpp

diff --git a/SYCL/Complex/sycl_complex_marray_math_test.cpp b/SYCL/Complex/sycl_complex_marray_math_test.cpp
new file mode 100644
index 0000000000..3c19bacfc5
--- /dev/null
+++ b/SYCL/Complex/sycl_complex_marray_math_test.cpp
@@ -0,0 +1,486 @@
+// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -fsycl-device-code-split=per_kernel %s -o %t.out
+// RUN: %CPU_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+// RUN: %ACC_RUN_PLACEHOLDER %t.out
+
+#include "sycl_complex_marray_test_cases.hpp"
+
+#define TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(func_name)                           \
+  template <typename T> struct test_##func_name {                              \
+    bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,    \
+                    const std::vector<cmplx<double>> &ref = {},                \
+                    bool use_ref = false) {                                    \
+      bool pass = true;                                                        \
+      using X = typename std::conditional<std::is_same<T, sycl::half>::value,  \
+                                          float, T>::type;                     \
+                                                                               \
+      /* std::complex test cases */                                            \
+      sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;            \
+      /* sycl::complex test cases */                                           \
+      sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;  \
+                                                                               \
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {               \
+        std_in[i] = init_std_complex<T>(init[i].re, init[i].im);               \
+        cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),      \
+                                              static_cast<T>(init[i].im)};     \
+      }                                                                        \
+                                                                               \
+      sycl::marray<std::complex<T>, DEFAULT_TEST_CASE_SIZE> std_out{};         \
+      auto *cplx_out = sycl::malloc_shared<                                    \
+          sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE>>(1,   \
+                                                                          Q);  \
+                                                                               \
+      /* Get std::complex output */                                            \
+      if (use_ref) {                                                           \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          T re = static_cast<T>(ref[i].re);                                    \
+          T im = static_cast<T>(ref[i].im);                                    \
+          std_out[i] = std::complex<T>{re, im};                                \
+        }                                                                      \
+      } else {                                                                 \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          std_out[i] = std::func_name(std_in[i]);                              \
+        }                                                                      \
+      }                                                                        \
+                                                                               \
+      /* Check cplx::complex output from device */                             \
+      Q.single_task([=]() {                                                    \
+         *cplx_out = experimental::func_name(cplx_in);                         \
+       }).wait();                                                              \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ true);           \
+                                                                               \
+      /* Check cplx::complex output from host */                               \
+      *cplx_out = experimental::func_name(cplx_in);                            \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ false);          \
+                                                                               \
+      sycl::free(cplx_out, Q);                                                 \
+                                                                               \
+      return pass;                                                             \
+    }                                                                          \
+  };
+
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(acos)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(asin)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(atan)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(acosh)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(asinh)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(atanh)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(conj)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(cos)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(cosh)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(exp)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(log)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(log10)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(proj)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(sin)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(sinh)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(sqrt)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(tan)
+TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT(tanh)
+
+#undef TEST_MATH_MARRAY_CPLX_IN_CPLX_OUT
+
+#define TEST_MATH_MARRAY_CPLX_IN_DECI_OUT(func_name)                           \
+  template <typename T> struct test_##func_name {                              \
+    bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,    \
+                    const std::vector<cmplx<double>> &ref = {},                \
+                    bool use_ref = false) {                                    \
+      bool pass = true;                                                        \
+                                                                               \
+      using X = typename std::conditional<std::is_same<T, sycl::half>::value,  \
+                                          float, T>::type;                     \
+                                                                               \
+      /* std::complex test cases */                                            \
+      sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;            \
+      /* sycl::complex test cases */                                           \
+      sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;  \
+                                                                               \
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {               \
+        std_in[i] = init_std_complex<T>(init[i].re, init[i].im);               \
+        cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),      \
+                                              static_cast<T>(init[i].im)};     \
+      }                                                                        \
+                                                                               \
+      sycl::marray<T, DEFAULT_TEST_CASE_SIZE> std_out{};                       \
+      auto *cplx_out =                                                         \
+          sycl::malloc_shared<sycl::marray<T, DEFAULT_TEST_CASE_SIZE>>(1, Q);  \
+                                                                               \
+      /* Get std::complex output */                                            \
+      if (use_ref) {                                                           \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          std_out[i] = static_cast<T>(ref[i].re);                              \
+        }                                                                      \
+      } else {                                                                 \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          std_out[i] = std::func_name(std_in[i]);                              \
+        }                                                                      \
+      }                                                                        \
+                                                                               \
+      /* Check cplx::complex output from device */                             \
+      Q.single_task([=]() {                                                    \
+         *cplx_out = experimental::func_name(cplx_in);                         \
+       }).wait();                                                              \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ true);           \
+                                                                               \
+      /* Check cplx::complex output from host */                               \
+      *cplx_out = experimental::func_name(cplx_in);                            \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ false);          \
+                                                                               \
+      sycl::free(cplx_out, Q);                                                 \
+                                                                               \
+      return pass;                                                             \
+    }                                                                          \
+  };
+
+TEST_MATH_MARRAY_CPLX_IN_DECI_OUT(abs)
+TEST_MATH_MARRAY_CPLX_IN_DECI_OUT(arg)
+TEST_MATH_MARRAY_CPLX_IN_DECI_OUT(norm)
+
+#undef TEST_MARRAY_CPLX_IN_DECI_OUT
+
+template <typename T> struct test_polar {
+  bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,
+                  const std::vector<cmplx<double>> &ref = {},
+                  bool use_ref = false) {
+    bool pass = true;
+
+    /* test cases */
+    sycl::marray<T, POLAR_TEST_CASE_SIZE> rho;
+    sycl::marray<T, POLAR_TEST_CASE_SIZE> theta;
+    for (std::size_t i = 0; i < POLAR_TEST_CASE_SIZE; ++i) {
+      rho[i] = static_cast<T>(init[i].re);
+      theta[i] = static_cast<T>(init[i].im);
+    }
+
+    sycl::marray<std::complex<T>, POLAR_TEST_CASE_SIZE> std_out{};
+    auto *cplx_out = sycl::malloc_shared<
+        sycl::marray<experimental::complex<T>, POLAR_TEST_CASE_SIZE>>(1, Q);
+
+    /* Get std::complex output */
+    if (use_ref) {
+      for (std::size_t i = 0; i < POLAR_TEST_CASE_SIZE; ++i) {
+        T re = static_cast<T>(ref[i].re);
+        T im = static_cast<T>(ref[i].im);
+        std_out[i] = std::complex<T>{re, im};
+      }
+    } else {
+      for (std::size_t i = 0; i < POLAR_TEST_CASE_SIZE; ++i) {
+        std_out[i] = std::polar(rho[i], theta[i]);
+      }
+    }
+
+    /* Check cplx::complex output from device */
+    Q.single_task([=]() {
+       *cplx_out = experimental::polar(rho, theta);
+     }).wait();
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ true);
+
+    /* Check cplx::complex output from host */
+    *cplx_out = experimental::polar(rho, theta);
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ false);
+
+    sycl::free(cplx_out, Q);
+
+    return pass;
+  }
+};
+
+template <typename T> struct test_pow_cplx_cplx {
+  bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,
+                  const std::vector<cmplx<double>> &ref = {},
+                  bool use_ref = false) {
+    bool pass = true;
+
+    using X = typename std::conditional<std::is_same<T, sycl::half>::value,
+                                        float, T>::type;
+
+    /* std::complex test cases */
+    sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;
+    /* sycl::complex test cases */
+    sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;
+
+    for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+      std_in[i] = init_std_complex<T>(init[i].re, init[i].im);
+      cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),
+                                            static_cast<T>(init[i].im)};
+    }
+
+    sycl::marray<std::complex<T>, DEFAULT_TEST_CASE_SIZE> std_out{};
+    auto *cplx_out = sycl::malloc_shared<
+        sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE>>(1, Q);
+
+    /* Get std::complex output */
+    if (use_ref) {
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+        T re = static_cast<T>(ref[i].re);
+        T im = static_cast<T>(ref[i].im);
+        std_out[i] = std::complex<T>{re, im};
+      }
+    } else {
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+        std_out[i] = std::pow(std_in[i], std_in[i]);
+      }
+    }
+
+    /* Check cplx::complex output from device */
+    Q.single_task([=]() {
+       *cplx_out = experimental::pow(cplx_in, cplx_in);
+     }).wait();
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ true);
+
+    /* Check cplx::complex output from host */
+    *cplx_out = experimental::pow(cplx_in, cplx_in);
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ false);
+
+    sycl::free(cplx_out, Q);
+
+    return pass;
+  }
+};
+
+template <typename T> struct test_pow_cplx_deci {
+  bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,
+                  const std::vector<cmplx<double>> &ref = {},
+                  bool use_ref = false) {
+    bool pass = true;
+
+    using X = typename std::conditional<std::is_same<T, sycl::half>::value,
+                                        float, T>::type;
+
+    /* std::complex test cases */
+    sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;
+    /* sycl::complex test cases */
+    sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;
+
+    for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+      std_in[i] = init_std_complex<T>(init[i].re, init[i].im);
+      cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),
+                                            static_cast<T>(init[i].im)};
+    }
+
+    sycl::marray<std::complex<T>, DEFAULT_TEST_CASE_SIZE> std_out{};
+    auto *cplx_out = sycl::malloc_shared<
+        sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE>>(1, Q);
+
+    /* Get std::complex output */
+    if (use_ref) {
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+        T re = static_cast<T>(ref[i].re);
+        T im = static_cast<T>(ref[i].im);
+        std_out[i] = std::complex<T>{re, im};
+      }
+    } else {
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+        std_out[i] = std::pow(std_in[i], std_in[i].real());
+      }
+    }
+
+    /* Check cplx::complex output from device */
+    Q.single_task([=]() {
+       *cplx_out = experimental::pow(cplx_in, cplx_in.real());
+     }).wait();
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ true);
+
+    /* Check cplx::complex output from host */
+    *cplx_out = experimental::pow(cplx_in, cplx_in.real());
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ false);
+
+    sycl::free(cplx_out, Q);
+
+    return pass;
+  }
+};
+
+template <typename T> struct test_pow_deci_cplx {
+  bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,
+                  const std::vector<cmplx<double>> &ref = {},
+                  bool use_ref = false) {
+    bool pass = true;
+
+    using X = typename std::conditional<std::is_same<T, sycl::half>::value,
+                                        float, T>::type;
+
+    /* std::complex test cases */
+    sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;
+    /* sycl::complex test cases */
+    sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;
+
+    for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+      std_in[i] = init_std_complex<T>(init[i].re, init[i].im);
+      cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),
+                                            static_cast<T>(init[i].im)};
+    }
+
+    sycl::marray<std::complex<T>, DEFAULT_TEST_CASE_SIZE> std_out{};
+    auto *cplx_out = sycl::malloc_shared<
+        sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE>>(1, Q);
+
+    /* Get std::complex output */
+    if (use_ref) {
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+        T re = static_cast<T>(ref[i].re);
+        T im = static_cast<T>(ref[i].im);
+        std_out[i] = std::complex<T>{re, im};
+      }
+    } else {
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {
+        std_out[i] = std::pow(std_in[i].real(), std_in[i]);
+      }
+    }
+
+    /* Check cplx::complex output from device */
+    Q.single_task([=]() {
+       *cplx_out = experimental::pow(cplx_in.real(), cplx_in);
+     }).wait();
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ true);
+
+    /* Check cplx::complex output from host */
+    *cplx_out = experimental::pow(cplx_in.real(), cplx_in);
+    pass &= check_results(*cplx_out, std_out, /*is_device*/ false);
+
+    sycl::free(cplx_out, Q);
+
+    return pass;
+  }
+};
+
+int main() {
+  sycl::queue Q;
+
+  bool test_passes = true;
+
+  /* Test complex in, complex out functions */
+
+  {
+    marray_cplx_test_cases<test_acos> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_asin> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_atan> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_acosh> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_asinh> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_atanh> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_conj> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_cos> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_cosh> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_exp> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_log> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_log10> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_proj> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_sin> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_sinh> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_sqrt> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_tan> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_tanh> test;
+    test_passes &= test(Q);
+  }
+
+  /* Test complex in, decimal out functions */
+
+  {
+    marray_cplx_test_cases<test_abs> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_arg> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_norm> test;
+    test_passes &= test(Q);
+  }
+
+  /* Test polar function */
+
+  {
+    marray_cplx_test_cases<test_polar> test;
+    test_passes &= test(Q);
+  }
+
+  /* Test pow function */
+
+  {
+    marray_cplx_test_cases<test_pow_cplx_cplx> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_pow_cplx_deci> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_pow_deci_cplx> test;
+    test_passes &= test(Q);
+  }
+
+  return !test_passes;
+}

From 428daf33cc878e27d626cbacbbf429636e0a4667 Mon Sep 17 00:00:00 2001
From: Jefferson Le Quellec <jefferson.lequellec@codeplay.com>
Date: Tue, 14 Mar 2023 17:08:29 +0100
Subject: [PATCH 6/6] Add marray complex operator tests

---
 .../sycl_complex_marray_operator_test.cpp     | 253 ++++++++++++++++++
 1 file changed, 253 insertions(+)
 create mode 100644 SYCL/Complex/sycl_complex_marray_operator_test.cpp

diff --git a/SYCL/Complex/sycl_complex_marray_operator_test.cpp b/SYCL/Complex/sycl_complex_marray_operator_test.cpp
new file mode 100644
index 0000000000..c7dca3502e
--- /dev/null
+++ b/SYCL/Complex/sycl_complex_marray_operator_test.cpp
@@ -0,0 +1,253 @@
+// RUN: %clangxx -fsycl -fsycl-targets=%sycl_triple -fsycl-device-code-split=per_kernel %s -o %t.out
+// RUN: %CPU_RUN_PLACEHOLDER %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+// RUN: %ACC_RUN_PLACEHOLDER %t.out
+
+#include "sycl_complex_marray_test_cases.hpp"
+
+#define TEST_BASIC_OPERATOR(op_name, op)                                       \
+  template <typename T> struct test_##op_name {                                \
+    bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,    \
+                    const std::vector<cmplx<double>> &ref = {},                \
+                    bool use_ref = false) {                                    \
+      bool pass = true;                                                        \
+                                                                               \
+      using X = typename std::conditional<std::is_same<T, sycl::half>::value,  \
+                                          float, T>::type;                     \
+                                                                               \
+      /* std::complex test cases */                                            \
+      sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;            \
+      /* sycl::complex test cases */                                           \
+      sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;  \
+                                                                               \
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {               \
+        std_in[i] = init_std_complex<T>(init[i].re, init[i].im);               \
+        cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),      \
+                                              static_cast<T>(init[i].im)};     \
+      }                                                                        \
+                                                                               \
+      sycl::marray<std::complex<T>, DEFAULT_TEST_CASE_SIZE> std_out{};         \
+      auto *cplx_out = sycl::malloc_shared<                                    \
+          sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE>>(1,   \
+                                                                          Q);  \
+                                                                               \
+      /* Get std::complex output */                                            \
+      if (use_ref) {                                                           \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          T re = static_cast<T>(ref[i].re);                                    \
+          T im = static_cast<T>(ref[i].im);                                    \
+          std_out[i] = std::complex<T>{re, im};                                \
+        }                                                                      \
+      } else {                                                                 \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          std_out[i] = std_in[i] op std_in[i];                                 \
+        }                                                                      \
+      }                                                                        \
+                                                                               \
+      /* Check cplx::complex output from device */                             \
+      Q.single_task([=]() { *cplx_out = cplx_in op cplx_in; }).wait();         \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ true);           \
+                                                                               \
+      /* Check cplx::complex output from host */                               \
+      *cplx_out = cplx_in op cplx_in;                                          \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ false);          \
+                                                                               \
+      sycl::free(cplx_out, Q);                                                 \
+                                                                               \
+      return pass;                                                             \
+    }                                                                          \
+  };
+
+TEST_BASIC_OPERATOR(add, +)
+TEST_BASIC_OPERATOR(sub, -)
+TEST_BASIC_OPERATOR(mul, *)
+TEST_BASIC_OPERATOR(div, /)
+
+#undef TEST_BASIC_OPERATOR
+
+#define TEST_ASSIGN_OPERATOR(op_name, op)                                      \
+  template <typename T> struct test_##op_name {                                \
+    bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,    \
+                    const std::vector<cmplx<double>> &ref = {},                \
+                    bool use_ref = false) {                                    \
+      bool pass = true;                                                        \
+                                                                               \
+      using X = typename std::conditional<std::is_same<T, sycl::half>::value,  \
+                                          float, T>::type;                     \
+                                                                               \
+      /* std::complex test cases */                                            \
+      sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;            \
+      /* sycl::complex test cases */                                           \
+      sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;  \
+                                                                               \
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {               \
+        std_in[i] = init_std_complex<T>(init[i].re, init[i].im);               \
+        cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),      \
+                                              static_cast<T>(init[i].im)};     \
+      }                                                                        \
+                                                                               \
+      sycl::marray<std::complex<T>, DEFAULT_TEST_CASE_SIZE> std_out{};         \
+      auto *cplx_out = sycl::malloc_shared<                                    \
+          sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE>>(1,   \
+                                                                          Q);  \
+                                                                               \
+      std_out = convert_marray<T, X>(std_in);                                  \
+      *cplx_out = cplx_in;                                                     \
+                                                                               \
+      /* Get std::complex output */                                            \
+      if (use_ref) {                                                           \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          T re = static_cast<T>(ref[i].re);                                    \
+          T im = static_cast<T>(ref[i].im);                                    \
+          std_out[i] = std::complex<T>{re, im};                                \
+        }                                                                      \
+      } else {                                                                 \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          std_out[i] op std_in[i];                                             \
+        }                                                                      \
+      }                                                                        \
+                                                                               \
+      /* Check cplx::complex output from device */                             \
+      Q.single_task([=]() { *cplx_out op cplx_in; }).wait();                   \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ true);           \
+                                                                               \
+      *cplx_out = cplx_in;                                                     \
+                                                                               \
+      /* Check cplx::complex output from host */                               \
+      *cplx_out op cplx_in;                                                    \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ false);          \
+                                                                               \
+      sycl::free(cplx_out, Q);                                                 \
+                                                                               \
+      return pass;                                                             \
+    }                                                                          \
+  };
+
+TEST_ASSIGN_OPERATOR(assign_add, +=)
+TEST_ASSIGN_OPERATOR(assign_sub, -=)
+TEST_ASSIGN_OPERATOR(assign_mul, *=)
+TEST_ASSIGN_OPERATOR(assign_div, /=)
+
+#undef TEST_ASSIGN_OPERATOR
+
+#define TEST_UNARY_OPERATOR(op_name, op)                                       \
+  template <typename T> struct test_##op_name {                                \
+    bool operator()(sycl::queue &Q, const std::vector<cmplx<double>> &init,    \
+                    const std::vector<cmplx<double>> &ref = {},                \
+                    bool use_ref = false) {                                    \
+      bool pass = true;                                                        \
+                                                                               \
+      using X = typename std::conditional<std::is_same<T, sycl::half>::value,  \
+                                          float, T>::type;                     \
+                                                                               \
+      /* std::complex test cases */                                            \
+      sycl::marray<std::complex<X>, DEFAULT_TEST_CASE_SIZE> std_in;            \
+      /* sycl::complex test cases */                                           \
+      sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE> cplx_in;  \
+                                                                               \
+      for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {               \
+        std_in[i] = init_std_complex<T>(init[i].re, init[i].im);               \
+        cplx_in[i] = experimental::complex<T>{static_cast<T>(init[i].re),      \
+                                              static_cast<T>(init[i].im)};     \
+      }                                                                        \
+                                                                               \
+      sycl::marray<std::complex<T>, DEFAULT_TEST_CASE_SIZE> std_out{};         \
+      auto *cplx_out = sycl::malloc_shared<                                    \
+          sycl::marray<experimental::complex<T>, DEFAULT_TEST_CASE_SIZE>>(1,   \
+                                                                          Q);  \
+                                                                               \
+      /* Get std::complex output */                                            \
+      if (use_ref) {                                                           \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          T re = static_cast<T>(ref[i].re);                                    \
+          T im = static_cast<T>(ref[i].im);                                    \
+          std_out[i] = std::complex<T>{re, im};                                \
+        }                                                                      \
+      } else {                                                                 \
+        for (std::size_t i = 0; i < DEFAULT_TEST_CASE_SIZE; ++i) {             \
+          std_out[i] = op std_in[i];                                           \
+        }                                                                      \
+      }                                                                        \
+                                                                               \
+      /* Check cplx::complex output from device */                             \
+      Q.single_task([=]() { *cplx_out = op cplx_in; }).wait();                 \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ true);           \
+                                                                               \
+      /* Check cplx::complex output from host */                               \
+      *cplx_out = op cplx_in;                                                  \
+      pass &= check_results(*cplx_out, std_out, /*is_device*/ false);          \
+                                                                               \
+      sycl::free(cplx_out, Q);                                                 \
+                                                                               \
+      return pass;                                                             \
+    }                                                                          \
+  };
+
+TEST_UNARY_OPERATOR(unary_add, +)
+TEST_UNARY_OPERATOR(unary_sub, -)
+
+#undef TEST_UNARY_OPERATOR
+
+int main() {
+  sycl::queue Q;
+
+  bool test_passes = true;
+
+  /* Test basic arithmetic operator */
+
+  {
+    marray_cplx_test_cases<test_add> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_sub> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_mul> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_div> test;
+    test_passes &= test(Q);
+  }
+
+  /* Test assign arithmetic operator */
+
+  {
+    marray_cplx_test_cases<test_assign_add> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_assign_sub> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_assign_mul> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_assign_div> test;
+    test_passes &= test(Q);
+  }
+
+  /* Test unary operator */
+
+  {
+    marray_cplx_test_cases<test_unary_add> test;
+    test_passes &= test(Q);
+  }
+
+  {
+    marray_cplx_test_cases<test_unary_sub> test;
+    test_passes &= test(Q);
+  }
+
+  return !test_passes;
+}