From 910e66f29ea58a395e82e9627a437e8b05f04ce1 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 25 Apr 2022 15:27:11 +0100 Subject: [PATCH 1/9] Add half non-assign math operators --- sycl/include/CL/sycl/half_type.hpp | 28 ++++++++++++++++++++++++++++ sycl/source/detail/builtins_math.cpp | 2 +- 2 files changed, 29 insertions(+), 1 deletion(-) diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index 3c436eacdfbd4..c9faa654cb5fe 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -394,6 +394,34 @@ class half { half r = *this; return -r; } + +// Operator +, -, *, / +#define OP(op, op1) \ + friend half operator op(const half &lhs, const half &rhs) { \ + half rtn = lhs; \ + rtn op1 rhs; \ + return rtn; \ + } \ + template ::value>> \ + friend half operator op(const half &lhs, const T &rhs) { \ + half rtn = lhs; \ + rtn op1 rhs; \ + return rtn; \ + } \ + template ::value>> \ + friend half operator op(const T &lhs, const half &rhs) { \ + half rtn = rhs; \ + rtn op1 lhs; \ + return rtn; \ + } + OP(+, +=) + OP(-, -=) + OP(*, *=) + OP(/, /=) +#undef OP + // Operator float __SYCL_CONSTEXPR_HALF operator float() const { return static_cast(Data); diff --git a/sycl/source/detail/builtins_math.cpp b/sycl/source/detail/builtins_math.cpp index 5a78d6cb80a5b..4737035eff628 100644 --- a/sycl/source/detail/builtins_math.cpp +++ b/sycl/source/detail/builtins_math.cpp @@ -42,7 +42,7 @@ template inline T __cospi(T x) { return std::cos(M_PI * x); } template T inline __fract(T x, T *iptr) { T f = std::floor(x); *(iptr) = f; - return std::fmin(x - f, nextafter(T(1.0), T(0.0))); + return std::fmin(x - f, std::nextafter(T(1.0), T(0.0))); } template inline T __lgamma_r(T x, s::cl_int *signp) { From 0595900838dfbfa55892fd10782f268c6455a089 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Tue, 26 Apr 2022 11:33:08 +0100 Subject: [PATCH 2/9] Make half operations+-*/ constexpr --- sycl/include/CL/sycl/half_type.hpp | 9 ++++++--- 1 file changed, 6 insertions(+), 3 deletions(-) diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index c9faa654cb5fe..271a55e7b8e78 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -397,21 +397,24 @@ class half { // Operator +, -, *, / #define OP(op, op1) \ - friend half operator op(const half &lhs, const half &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ + const half &rhs) { \ half rtn = lhs; \ rtn op1 rhs; \ return rtn; \ } \ template ::value>> \ - friend half operator op(const half &lhs, const T &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ + const T &rhs) { \ half rtn = lhs; \ rtn op1 rhs; \ return rtn; \ } \ template ::value>> \ - friend half operator op(const T &lhs, const half &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op(const T &lhs, \ + const half &rhs) { \ half rtn = rhs; \ rtn op1 lhs; \ return rtn; \ From b000431beb6a0507058c65b7d8656354dbabbb97 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Wed, 27 Apr 2022 11:01:24 +0100 Subject: [PATCH 3/9] Add test and update based on feedback --- sycl/include/CL/sycl/half_type.hpp | 56 +++++++++++++++++-- sycl/test/type_traits/half_operator_types.cpp | 33 +++++++++++ 2 files changed, 83 insertions(+), 6 deletions(-) create mode 100644 sycl/test/type_traits/half_operator_types.cpp diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index 271a55e7b8e78..4153c0ad1ba69 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -403,17 +403,61 @@ class half { rtn op1 rhs; \ return rtn; \ } \ - template ::value>> \ + __SYCL_CONSTEXPR_HALF friend double operator op(const half &lhs, \ + const double &rhs) { \ + double rtn = lhs; \ + rtn op1 rhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend double operator op(const double &lhs, \ + const half &rhs) { \ + double rtn = rhs; \ + rtn op1 lhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend float operator op(const half &lhs, \ + const float &rhs) { \ + float rtn = lhs; \ + rtn op1 rhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend float operator op(const float &lhs, \ + const half &rhs) { \ + float rtn = rhs; \ + rtn op1 lhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ + const int &rhs) { \ + half rtn = lhs; \ + rtn op1 rhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const int &lhs, \ + const half &rhs) { \ + half rtn = rhs; \ + rtn op1 lhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ + const long &rhs) { \ + half rtn = lhs; \ + rtn op1 rhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const long &lhs, \ + const half &rhs) { \ + half rtn = rhs; \ + rtn op1 lhs; \ + return rtn; \ + } \ __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ - const T &rhs) { \ + const long long &rhs) { \ half rtn = lhs; \ rtn op1 rhs; \ return rtn; \ } \ - template ::value>> \ - __SYCL_CONSTEXPR_HALF friend half operator op(const T &lhs, \ + __SYCL_CONSTEXPR_HALF friend half operator op(const long long &lhs, \ const half &rhs) { \ half rtn = rhs; \ rtn op1 lhs; \ diff --git a/sycl/test/type_traits/half_operator_types.cpp b/sycl/test/type_traits/half_operator_types.cpp new file mode 100644 index 0000000000000..e422dc115c103 --- /dev/null +++ b/sycl/test/type_traits/half_operator_types.cpp @@ -0,0 +1,33 @@ +// RUN: %clangxx -fsycl %s -o %t.out +//==-------------- type_traits.cpp - SYCL type_traits test -----------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// + +#include + +template +void check_half_operator_types() { + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); +} + +int main() { + check_half_operator_types(); + check_half_operator_types(); + check_half_operator_types(); + check_half_operator_types(); + check_half_operator_types(); + + check_half_operator_types(); + check_half_operator_types(); + check_half_operator_types(); + check_half_operator_types(); + check_half_operator_types(); + check_half_operator_types(); +} From b3c7c400bf5b18c3beb50efda90347838e799cd8 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Wed, 27 Apr 2022 11:12:20 +0100 Subject: [PATCH 4/9] Change op1 to op_eq --- sycl/include/CL/sycl/half_type.hpp | 24 ++++++++++++------------ 1 file changed, 12 insertions(+), 12 deletions(-) diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index 4153c0ad1ba69..84b572bbb3715 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -396,71 +396,71 @@ class half { } // Operator +, -, *, / -#define OP(op, op1) \ +#define OP(op, op_eq) \ __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ const half &rhs) { \ half rtn = lhs; \ - rtn op1 rhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend double operator op(const half &lhs, \ const double &rhs) { \ double rtn = lhs; \ - rtn op1 rhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend double operator op(const double &lhs, \ const half &rhs) { \ double rtn = rhs; \ - rtn op1 lhs; \ + rtn op_eq lhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend float operator op(const half &lhs, \ const float &rhs) { \ float rtn = lhs; \ - rtn op1 rhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend float operator op(const float &lhs, \ const half &rhs) { \ float rtn = rhs; \ - rtn op1 lhs; \ + rtn op_eq lhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ const int &rhs) { \ half rtn = lhs; \ - rtn op1 rhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const int &lhs, \ const half &rhs) { \ half rtn = rhs; \ - rtn op1 lhs; \ + rtn op_eq lhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ const long &rhs) { \ half rtn = lhs; \ - rtn op1 rhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const long &lhs, \ const half &rhs) { \ half rtn = rhs; \ - rtn op1 lhs; \ + rtn op_eq lhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ const long long &rhs) { \ half rtn = lhs; \ - rtn op1 rhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const long long &lhs, \ const half &rhs) { \ half rtn = rhs; \ - rtn op1 lhs; \ + rtn op_eq lhs; \ return rtn; \ } OP(+, +=) From e0401ae99b19dd34bfd8750248f110e4b1ba1283 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Wed, 27 Apr 2022 17:08:57 +0100 Subject: [PATCH 5/9] add logical half operators --- sycl/include/CL/sycl/half_type.hpp | 56 +++++++++++++++++++ sycl/test/type_traits/half_operator_types.cpp | 55 ++++++++++++------ 2 files changed, 94 insertions(+), 17 deletions(-) diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index 84b572bbb3715..c94dab3c5187a 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -467,6 +467,62 @@ class half { OP(-, -=) OP(*, *=) OP(/, /=) + +#undef OP + +// Operator ==, !=, <, >, <=, >= +#define OP(op) \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const half &rhs) { \ + return lhs.Data op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const double &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const double &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const float &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const float &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const int &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const int &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const long &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const long &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const long long &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const long long &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ + } + OP(==) + OP(!=) + OP(<) + OP(>) + OP(<=) + OP(>=) + #undef OP // Operator float diff --git a/sycl/test/type_traits/half_operator_types.cpp b/sycl/test/type_traits/half_operator_types.cpp index e422dc115c103..461d200b4ce56 100644 --- a/sycl/test/type_traits/half_operator_types.cpp +++ b/sycl/test/type_traits/half_operator_types.cpp @@ -9,25 +9,46 @@ #include -template -void check_half_operator_types() { - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); +template void check_half_math_operator_types() { + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); + + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); +} + +template void check_half_logical_operator_types() { + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same sycl::half(1)), bool>::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); + + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same T1(1)), bool>::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); + static_assert(std::is_same::value); } int main() { - check_half_operator_types(); - check_half_operator_types(); - check_half_operator_types(); - check_half_operator_types(); - check_half_operator_types(); + check_half_math_operator_types(); + check_half_math_operator_types(); + check_half_math_operator_types(); + check_half_math_operator_types(); + check_half_math_operator_types(); + check_half_math_operator_types(); - check_half_operator_types(); - check_half_operator_types(); - check_half_operator_types(); - check_half_operator_types(); - check_half_operator_types(); - check_half_operator_types(); + check_half_logical_operator_types(); + check_half_logical_operator_types(); + check_half_logical_operator_types(); + check_half_logical_operator_types(); + check_half_logical_operator_types(); + check_half_logical_operator_types(); } From fb7f65cb8d23f9020a13c08f0733f79444e3b5c8 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Fri, 29 Apr 2022 09:44:39 +0100 Subject: [PATCH 6/9] add unsigned types to half operators --- sycl/include/CL/sycl/half_type.hpp | 60 ++++++++++++++++++++++++++++++ 1 file changed, 60 insertions(+) diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index c94dab3c5187a..d83486d3efe44 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -462,6 +462,42 @@ class half { half rtn = rhs; \ rtn op_eq lhs; \ return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ + const unsigned int &rhs) { \ + half rtn = lhs; \ + rtn op_eq rhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned int &lhs, \ + const half &rhs) { \ + half rtn = rhs; \ + rtn op_eq lhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ + const unsigned long &rhs) { \ + half rtn = lhs; \ + rtn op_eq rhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long &lhs, \ + const half &rhs) { \ + half rtn = rhs; \ + rtn op_eq lhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ + const unsigned long long &rhs) { \ + half rtn = lhs; \ + rtn op_eq rhs; \ + return rtn; \ + } \ + __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long long &lhs, \ + const half &rhs) { \ + half rtn = rhs; \ + rtn op_eq lhs; \ + return rtn; \ } OP(+, +=) OP(-, -=) @@ -515,6 +551,30 @@ class half { __SYCL_CONSTEXPR_HALF friend bool operator op(const long long &lhs, \ const half &rhs) { \ return lhs op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const unsigned int &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned int &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const unsigned long &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned long &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ + const unsigned long long &rhs) { \ + return lhs.Data op rhs; \ + } \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned long long &lhs, \ + const half &rhs) { \ + return lhs op rhs.Data; \ } OP(==) OP(!=) From b05e58a0897050c02f3056dc7d7a47019f9bfa65 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Mon, 2 May 2022 10:38:57 +0100 Subject: [PATCH 7/9] clang format --- sycl/include/CL/sycl/half_type.hpp | 24 ++++++++++++------------ 1 file changed, 12 insertions(+), 12 deletions(-) diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index d83486d3efe44..7caa4b9198146 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -481,19 +481,19 @@ class half { rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long &lhs, \ + __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long &lhs, \ const half &rhs) { \ half rtn = rhs; \ rtn op_eq lhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ - const unsigned long long &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op( \ + const half &lhs, const unsigned long long &rhs) { \ half rtn = lhs; \ rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long long &lhs, \ + __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long long &lhs, \ const half &rhs) { \ half rtn = rhs; \ rtn op_eq lhs; \ @@ -551,28 +551,28 @@ class half { __SYCL_CONSTEXPR_HALF friend bool operator op(const long long &lhs, \ const half &rhs) { \ return lhs op rhs.Data; \ - } \ + } \ __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ - const unsigned int &rhs) { \ + const unsigned int &rhs) { \ return lhs.Data op rhs; \ } \ - __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned int &lhs, \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned int &lhs, \ const half &rhs) { \ return lhs op rhs.Data; \ } \ __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ - const unsigned long &rhs) { \ + const unsigned long &rhs) { \ return lhs.Data op rhs; \ } \ - __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned long &lhs, \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned long &lhs, \ const half &rhs) { \ return lhs op rhs.Data; \ } \ - __SYCL_CONSTEXPR_HALF friend bool operator op(const half &lhs, \ - const unsigned long long &rhs) { \ + __SYCL_CONSTEXPR_HALF friend bool operator op( \ + const half &lhs, const unsigned long long &rhs) { \ return lhs.Data op rhs; \ } \ - __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned long long &lhs, \ + __SYCL_CONSTEXPR_HALF friend bool operator op(const unsigned long long &lhs, \ const half &rhs) { \ return lhs op rhs.Data; \ } From 45d93e69ae9a6d3acd9f210e0218d3f780a659a8 Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Tue, 3 May 2022 11:25:02 +0100 Subject: [PATCH 8/9] fix neg, div mistake. Add device-code testing --- sycl/include/CL/sycl/half_type.hpp | 85 ++++++------- sycl/test/type_traits/half_operator_types.cpp | 117 ++++++++++++------ 2 files changed, 120 insertions(+), 82 deletions(-) diff --git a/sycl/include/CL/sycl/half_type.hpp b/sycl/include/CL/sycl/half_type.hpp index 7caa4b9198146..5fc1559934ab5 100644 --- a/sycl/include/CL/sycl/half_type.hpp +++ b/sycl/include/CL/sycl/half_type.hpp @@ -386,81 +386,76 @@ class half { operator--(); return ret; } - constexpr half &operator-() { - Data = -Data; - return *this; - } - constexpr half operator-() const { - half r = *this; - return -r; + __SYCL_CONSTEXPR_HALF friend half operator-(const half other) { + return half(-other.Data); } // Operator +, -, *, / #define OP(op, op_eq) \ - __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ - const half &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half lhs, \ + const half rhs) { \ half rtn = lhs; \ rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend double operator op(const half &lhs, \ - const double &rhs) { \ + __SYCL_CONSTEXPR_HALF friend double operator op(const half lhs, \ + const double rhs) { \ double rtn = lhs; \ rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend double operator op(const double &lhs, \ - const half &rhs) { \ - double rtn = rhs; \ - rtn op_eq lhs; \ + __SYCL_CONSTEXPR_HALF friend double operator op(const double lhs, \ + const half rhs) { \ + double rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend float operator op(const half &lhs, \ - const float &rhs) { \ + __SYCL_CONSTEXPR_HALF friend float operator op(const half lhs, \ + const float rhs) { \ float rtn = lhs; \ rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend float operator op(const float &lhs, \ - const half &rhs) { \ - float rtn = rhs; \ - rtn op_eq lhs; \ + __SYCL_CONSTEXPR_HALF friend float operator op(const float lhs, \ + const half rhs) { \ + float rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ - const int &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half lhs, \ + const int rhs) { \ half rtn = lhs; \ rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const int &lhs, \ - const half &rhs) { \ - half rtn = rhs; \ - rtn op_eq lhs; \ + __SYCL_CONSTEXPR_HALF friend half operator op(const int lhs, \ + const half rhs) { \ + half rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ - const long &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half lhs, \ + const long rhs) { \ half rtn = lhs; \ rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const long &lhs, \ - const half &rhs) { \ - half rtn = rhs; \ - rtn op_eq lhs; \ + __SYCL_CONSTEXPR_HALF friend half operator op(const long lhs, \ + const half rhs) { \ + half rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ - const long long &rhs) { \ + __SYCL_CONSTEXPR_HALF friend half operator op(const half lhs, \ + const long long rhs) { \ half rtn = lhs; \ rtn op_eq rhs; \ return rtn; \ } \ - __SYCL_CONSTEXPR_HALF friend half operator op(const long long &lhs, \ - const half &rhs) { \ - half rtn = rhs; \ - rtn op_eq lhs; \ + __SYCL_CONSTEXPR_HALF friend half operator op(const long long lhs, \ + const half rhs) { \ + half rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ @@ -471,8 +466,8 @@ class half { } \ __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned int &lhs, \ const half &rhs) { \ - half rtn = rhs; \ - rtn op_eq lhs; \ + half rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op(const half &lhs, \ @@ -483,8 +478,8 @@ class half { } \ __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long &lhs, \ const half &rhs) { \ - half rtn = rhs; \ - rtn op_eq lhs; \ + half rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } \ __SYCL_CONSTEXPR_HALF friend half operator op( \ @@ -495,8 +490,8 @@ class half { } \ __SYCL_CONSTEXPR_HALF friend half operator op(const unsigned long long &lhs, \ const half &rhs) { \ - half rtn = rhs; \ - rtn op_eq lhs; \ + half rtn = lhs; \ + rtn op_eq rhs; \ return rtn; \ } OP(+, +=) diff --git a/sycl/test/type_traits/half_operator_types.cpp b/sycl/test/type_traits/half_operator_types.cpp index 461d200b4ce56..2cee8f0b32d3e 100644 --- a/sycl/test/type_traits/half_operator_types.cpp +++ b/sycl/test/type_traits/half_operator_types.cpp @@ -9,46 +9,89 @@ #include -template void check_half_math_operator_types() { - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); - - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); +template +void check_half_math_operator_types(sycl::queue Queue) { + + // Test on host + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + + // Test on device + Queue.submit([&](sycl::handler &cgh) { + cgh.single_task([=] { + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + }); + }); } -template void check_half_logical_operator_types() { - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same sycl::half(1)), bool>::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); - - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same T1(1)), bool>::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); - static_assert(std::is_same::value); +template +void check_half_logical_operator_types(sycl::queue Queue) { + + // Test on host + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v sycl::half(1)), bool>); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v T1(1)), bool>); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + + // Test on device + Queue.submit([&](sycl::handler &cgh) { + cgh.single_task([=] { + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v sycl::half(1)), bool>); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v T1(1)), bool>); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + static_assert(std::is_same_v); + }); + }); } int main() { - check_half_math_operator_types(); - check_half_math_operator_types(); - check_half_math_operator_types(); - check_half_math_operator_types(); - check_half_math_operator_types(); - check_half_math_operator_types(); - - check_half_logical_operator_types(); - check_half_logical_operator_types(); - check_half_logical_operator_types(); - check_half_logical_operator_types(); - check_half_logical_operator_types(); - check_half_logical_operator_types(); + + sycl::queue Queue; + + check_half_math_operator_types(Queue); + check_half_math_operator_types(Queue); + check_half_math_operator_types(Queue); + check_half_math_operator_types(Queue); + check_half_math_operator_types(Queue); + check_half_math_operator_types(Queue); + + check_half_logical_operator_types(Queue); + check_half_logical_operator_types(Queue); + check_half_logical_operator_types(Queue); + check_half_logical_operator_types(Queue); + check_half_logical_operator_types(Queue); + check_half_logical_operator_types(Queue); } From b71bb8d6b5528d4860ff271c29d996e252258aae Mon Sep 17 00:00:00 2001 From: "aidan.belton" Date: Tue, 3 May 2022 18:03:07 +0100 Subject: [PATCH 9/9] update test based on feedback --- sycl/test/type_traits/half_operator_types.cpp | 105 +++++++++--------- 1 file changed, 55 insertions(+), 50 deletions(-) diff --git a/sycl/test/type_traits/half_operator_types.cpp b/sycl/test/type_traits/half_operator_types.cpp index 2cee8f0b32d3e..6d53356ea7de7 100644 --- a/sycl/test/type_traits/half_operator_types.cpp +++ b/sycl/test/type_traits/half_operator_types.cpp @@ -8,72 +8,77 @@ //===----------------------------------------------------------------------===// #include +using namespace std; + +template void math_operator_helper() { + static_assert( + is_same_v() + declval()), T_rtn>); + static_assert( + is_same_v() - declval()), T_rtn>); + static_assert( + is_same_v() * declval()), T_rtn>); + static_assert( + is_same_v() / declval()), T_rtn>); + + static_assert( + is_same_v() + declval()), T_rtn>); + static_assert( + is_same_v() - declval()), T_rtn>); + static_assert( + is_same_v() * declval()), T_rtn>); + static_assert( + is_same_v() / declval()), T_rtn>); +} + +template void logical_operator_helper() { + static_assert( + is_same_v() == declval()), bool>); + static_assert( + is_same_v() != declval()), bool>); + static_assert( + is_same_v() > declval()), bool>); + static_assert( + is_same_v() < declval()), bool>); + static_assert( + is_same_v() <= declval()), bool>); + static_assert( + is_same_v() >= declval()), bool>); + + static_assert( + is_same_v() == declval()), bool>); + static_assert( + is_same_v() != declval()), bool>); + static_assert( + is_same_v() > declval()), bool>); + static_assert( + is_same_v() < declval()), bool>); + static_assert( + is_same_v() <= declval()), bool>); + static_assert( + is_same_v() >= declval()), bool>); +} template -void check_half_math_operator_types(sycl::queue Queue) { +void check_half_math_operator_types(sycl::queue &Queue) { // Test on host - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); + math_operator_helper(); // Test on device Queue.submit([&](sycl::handler &cgh) { - cgh.single_task([=] { - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - }); + cgh.single_task([=] { math_operator_helper(); }); }); } template -void check_half_logical_operator_types(sycl::queue Queue) { +void check_half_logical_operator_types(sycl::queue &Queue) { // Test on host - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v sycl::half(1)), bool>); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v T1(1)), bool>); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); + logical_operator_helper(); // Test on device Queue.submit([&](sycl::handler &cgh) { - cgh.single_task([=] { - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v sycl::half(1)), bool>); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v T1(1)), bool>); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - static_assert(std::is_same_v); - }); + cgh.single_task([=] { logical_operator_helper(); }); }); }