[SYCL][ESIMD] Split test into sint and uint

Because we have possibility of trap values and compiler optimizations
in general C++17.

Author: vasilytric

SYCL/ESIMD/api/functional/operators/operator_bitwise_not.cpp renamed to SYCL/ESIMD/api/functional/operators/operator_bitwise_not.hpp

@@ -1,49 +1,43 @@
-//==------- operator_bitwise_not.cpp  - DPC++ ESIMD on-device test ---------==//
+//===-- operator_bitwise_not.hpp - Functions for tests on simd assignment
+//      operators. --------------------------------------------------------===//
 //
 // 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
 //
 //===----------------------------------------------------------------------===//
-// REQUIRES: gpu, level_zero
-// XREQUIRES: gpu
-// TODO gpu and level_zero in REQUIRES due to only this platforms supported yet.
-// The current "REQUIRES" should be replaced with "gpu" only as mentioned in
-// "XREQUIRES".
-// UNSUPPORTED: cuda, hip
-// RUN: %clangxx -fsycl %s -fsycl-device-code-split=per_kernel -o %t.out
-// RUN: %GPU_RUN_PLACEHOLDER %t.out
-//
-// Test for simd bitwise not operator.
-// The test creates source simd instance with reference data and invokes bitwise
-// not operator.
-// The test verifies that data from simd is not corrupted after calling bitwise
-// not operator, that bitwise not operator return type is as expected and
-// bitwise not operator result values is correct.
+///
+/// \file
+/// This file provides functions for tests on simd assignment operators.
+///
+//===----------------------------------------------------------------------===//
+
+#pragma once
 
 #include "../mutator.hpp"
 #include "../shared_element.hpp"
 #include "common.hpp"
 // For std::abs
 #include <cmath>
 
-using namespace sycl::ext::intel::experimental::esimd;
-using namespace esimd_test::api::functional;
+namespace esimd = sycl::ext::intel::esimd;
+
+namespace esimd_test::api::functional::operators {
 
-// Descriptor class for the case of calling bitwise not operator.
+  // Descriptor class for the case of calling bitwise not operator.
 struct bitwise_not_operator {
   static std::string get_description() { return "bitwise not"; }
 
   template <typename DataT, int NumElems>
   static bool call_operator(const DataT *const ref_data,
                             DataT *const source_simd_result,
                             DataT *const operator_result) {
-    auto simd_obj = simd<DataT, NumElems>();
+    auto simd_obj = esimd::simd<DataT, NumElems>();
     simd_obj.copy_from(ref_data);
     auto bitwise_not_result = ~simd_obj;
     simd_obj.copy_to(source_simd_result);
     bitwise_not_result.copy_to(operator_result);
-    return std::is_same_v<decltype(~simd_obj), simd<DataT, NumElems>>;
+    return std::is_same_v<decltype(~simd_obj), esimd::simd<DataT, NumElems>>;
   }
 };
 
@@ -169,21 +163,5 @@ template <typename TestCaseT, typename DataT, typename DimT> class run_test {
   }
 };
 
-int main(int, char **) {
-  sycl::queue queue(esimd_test::ESIMDSelector{},
-                    esimd_test::createExceptionHandler());
-
-  bool passed = true;
-
-  const auto uint_types = get_tested_types<tested_types::uint>();
-  const auto sint_types = get_tested_types<tested_types::sint>();
-  const auto all_dims = get_all_dimensions();
-
-  passed &= for_all_combinations<run_test, bitwise_not_operator>(
-      uint_types, all_dims, queue);
-  passed &= for_all_combinations<run_test, bitwise_not_operator>(
-      sint_types, all_dims, queue);
 
-  std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
-  return passed ? 0 : 1;
-}
+}

SYCL/ESIMD/api/functional/operators/operator_bitwise_not_sint.cpp

@@ -0,0 +1,49 @@
+//==------- operator_bitwise_not_sint.cpp  - DPC++ ESIMD on-device 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
+//
+//===----------------------------------------------------------------------===//
+// REQUIRES: gpu, level_zero
+// XREQUIRES: gpu
+// TODO gpu and level_zero in REQUIRES due to only this platforms supported yet.
+// The current "REQUIRES" should be replaced with "gpu" only as mentioned in
+// "XREQUIRES".
+// UNSUPPORTED: cuda, hip
+// RUN: %clangxx -fsycl %s -fsycl-device-code-split=per_kernel -o %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+//
+// Test for simd bitwise not operator.
+// The test creates source simd instance with reference data and invokes bitwise
+// not operator.
+// The test verifies that data from simd is not corrupted after calling bitwise
+// not operator, that bitwise not operator return type is as expected and
+// bitwise not operator result values is correct.
+//
+// IMPORTANT: there is possibility of trap values and compiler optimizations in
+// general C++17, the llvm itself states that signed integer types are
+// guaranteed to be two's complement:
+//  - https://llvm.org/docs/LangRef.html
+//  - https://bugs.llvm.org/show_bug.cgi?id=950
+
+#include "operator_bitwise_not.hpp"
+
+using namespace esimd_test::api::functional;
+
+int main(int, char **) {
+  sycl::queue queue(esimd_test::ESIMDSelector{},
+                    esimd_test::createExceptionHandler());
+
+  bool passed = true;
+
+  const auto sint_types = get_tested_types<tested_types::sint>();
+  const auto all_dims = get_all_dimensions();
+
+  // Running test for all sint types
+  passed &= for_all_combinations<run_test, bitwise_not_operator>(
+      sint_types, all_dims, queue);
+
+  std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
+  return passed ? 0 : 1;
+}

SYCL/ESIMD/api/functional/operators/operator_bitwise_not_uint.cpp

@@ -0,0 +1,44 @@
+//==------- operator_bitwise_not_uint.cpp  - DPC++ ESIMD on-device 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
+//
+//===----------------------------------------------------------------------===//
+// REQUIRES: gpu, level_zero
+// XREQUIRES: gpu
+// TODO gpu and level_zero in REQUIRES due to only this platforms supported yet.
+// The current "REQUIRES" should be replaced with "gpu" only as mentioned in
+// "XREQUIRES".
+// UNSUPPORTED: cuda, hip
+// RUN: %clangxx -fsycl %s -fsycl-device-code-split=per_kernel -o %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+//
+// Test for simd bitwise not operator.
+// The test creates source simd instance with reference data and invokes bitwise
+// not operator.
+// The test verifies that data from simd is not corrupted after calling bitwise
+// not operator, that bitwise not operator return type is as expected and
+// bitwise not operator result values is correct.
+
+#include "operator_bitwise_not.hpp"
+
+using namespace esimd_test::api::functional;
+
+int main(int, char **) {
+  sycl::queue queue(esimd_test::ESIMDSelector{},
+                    esimd_test::createExceptionHandler());
+
+  bool passed = true;
+
+  const auto uint_types = get_tested_types<tested_types::uint>();
+  const auto all_dims = get_all_dimensions();
+
+  // Running test for all uint types
+  passed &= for_all_combinations<operators::run_test,
+                                 operators::bitwise_not_operator>(
+      uint_types, all_dims, queue);
+
+  std::cout << (passed ? "=== Test passed\n" : "=== Test FAILED\n");
+  return passed ? 0 : 1;
+}