From ccf9155ab41f6e9187629583850b7bdde3fb76eb Mon Sep 17 00:00:00 2001 From: Harald van Dijk Date: Tue, 7 Nov 2023 17:56:27 +0000 Subject: [PATCH] Do not check for undefined behavior in abs. Unlike OpenCL abs, but like C and C++ abs, in SYCL 2020, taking the absolute value of a signed type produces a signed result and the behavior is undefined if the absolute value is not representable (i.e. if the input is the greatest negative value), so do not check that the SYCL results match the reference results for this case. --- util/math_reference.cpp | 2 +- util/math_reference.h | 17 ++++++++++++++--- 2 files changed, 15 insertions(+), 4 deletions(-) diff --git a/util/math_reference.cpp b/util/math_reference.cpp index c7ea1006d..67b34ef12 100644 --- a/util/math_reference.cpp +++ b/util/math_reference.cpp @@ -203,7 +203,7 @@ T mad_sat_signed_long(T x, T y, T z) { if ((x > 0 && y > 0) || (x < 0 && y < 0)) if (mul > 0 && std::abs(mul) > std::abs(x) && std::abs(mul) > std::abs(y)) return add_sat(mul, z); - else if (z < 0 && mul - std::numeric_limits::min() < abs(z)) + else if (z < 0 && mul - std::numeric_limits::min() < std::abs(z)) return std::numeric_limits::max() + z; else return std::numeric_limits::max(); diff --git a/util/math_reference.h b/util/math_reference.h index 162ddc5e1..d24ab754b 100644 --- a/util/math_reference.h +++ b/util/math_reference.h @@ -285,10 +285,21 @@ sycl::marray select(sycl::marray a, sycl::marray b, /* absolute value */ template -T abs(T x) { - return x < 0 ? -x : x; +sycl_cts::resultRef abs(T x) { + using U = typename std::make_unsigned::type; + T result = x < 0 ? T(-U(x)) : x; + return result < 0 ? sycl_cts::resultRef(0, true) : result; +} +template +sycl_cts::resultRef> abs(sycl::vec a) { + return sycl_cts::math::run_func_on_vector_result_ref( + [](T x) { return abs(x); }, a); +} +template +sycl_cts::resultRef> abs(sycl::marray a) { + return sycl_cts::math::run_func_on_marray_result_ref( + [](T x) { return abs(x); }, a); } -MAKE_VEC_AND_MARRAY_VERSIONS(abs) /* absolute difference */ template