Merge from 'sycl' to 'sycl-web'

Author: iclsrc

.github/workflows/clang-format.yml

@@ -4,7 +4,6 @@ on:
   pull_request:
     branches:
     - sycl
-    types: [open, edit, reopen, synchronize]
 
 jobs:
   build:
@@ -22,7 +21,7 @@ jobs:
 
     - name: Run clang-format for the patch
       run: |
-        git diff -U0 --no-color origin/${{github.base_ref}}..HEAD | ./clang/tools/clang-format/clang-format-diff.py -p1 -binary clang-format-9 > ./clang-format.patch
+        git diff -U0 --no-color origin/${{github.base_ref}}...HEAD | ./clang/tools/clang-format/clang-format-diff.py -p1 -binary clang-format-9 > ./clang-format.patch
 
     # Add patch with formatting fixes to CI job artifacts
     - uses: actions/upload-artifact@v1

buildbot/dependency.py

@@ -71,6 +71,7 @@ def do_dependency(args):
     install_dir = os.path.join(args.obj_dir, "install")
     cmake_cmd = ["cmake", "-G", "Ninja",
                  "-DCMAKE_INSTALL_PREFIX={}".format(install_dir),
+                 "-DOPENCL_ICD_LOADER_HEADERS_DIR={}".format(ocl_header_dir),
                  ".." ]
     subprocess.check_call(cmake_cmd, cwd=icd_build_dir)
 

clang/lib/Sema/CMakeLists.txt

@@ -9,6 +9,7 @@ if (MSVC)
   set_source_files_properties(SemaExprCXX.cpp PROPERTIES COMPILE_FLAGS /bigobj)
   set_source_files_properties(SemaSYCL.cpp PROPERTIES COMPILE_FLAGS /bigobj)
   set_source_files_properties(SemaTemplate.cpp PROPERTIES COMPILE_FLAGS /bigobj)
+  set_source_files_properties(SemaTemplateDeduction.cpp PROPERTIES COMPILE_FLAGS /bigobj)
 endif()
 
 clang_tablegen(OpenCLBuiltins.inc -gen-clang-opencl-builtins

sycl/include/CL/sycl/builtins.hpp

@@ -851,10 +851,6 @@ detail::enable_if_t<detail::is_ugeninteger<T>::value, T> sub_sat(T x,
   return __sycl_std::__invoke_u_sub_sat<T>(x, y);
 }
 
-// TODO delete when Intel CPU OpenCL runtime will be fixed
-// OpExtInst ... s_upsample -> _Z8upsampleij (now _Z8upsampleii)
-#define __invoke_s_upsample __invoke_u_upsample
-
 // ugeninteger16bit upsample (ugeninteger8bit hi, ugeninteger8bit lo)
 template <typename T>
 detail::enable_if_t<detail::is_ugeninteger8bit<T>::value,
@@ -909,8 +905,6 @@ upsample(T hi, T2 lo) __NOEXC {
   return __sycl_std::__invoke_s_upsample<detail::make_larger_t<T>>(hi, lo);
 }
 
-#undef __invoke_s_upsample
-
 // geninteger popcount (geninteger x)
 template <typename T>
 detail::enable_if_t<detail::is_geninteger<T>::value, T> popcount(T x) __NOEXC {

sycl/include/CL/sycl/half_type.hpp

@@ -18,6 +18,14 @@
 #include <iostream>
 #include <limits>
 
+#ifdef __SYCL_DEVICE_ONLY__
+// `constexpr` could work because the implicit conversion from `float` to
+// `_Float16` can be `constexpr`.
+#define __SYCL_CONSTEXPR_ON_DEVICE constexpr
+#else
+#define __SYCL_CONSTEXPR_ON_DEVICE
+#endif
+
 __SYCL_INLINE_NAMESPACE(cl) {
 namespace sycl {
 namespace detail {
@@ -130,7 +138,7 @@ class half {
   half(const half &) = default;
   half(half &&) = default;
 
-  half(const float &rhs) : Data(rhs) {}
+  __SYCL_CONSTEXPR_ON_DEVICE half(const float &rhs) : Data(rhs) {}
 
   half &operator=(const half &rhs) = default;
 
@@ -216,15 +224,6 @@ using half = cl::sycl::detail::half_impl::half;
 // Partial specialization of some functions in namespace `std`
 namespace std {
 
-#ifdef __SYCL_DEVICE_ONLY__
-// `constexpr` could work because the implicit conversion from `float` to
-// `_Float16` can be `constexpr`.
-#define CONSTEXPR_QUALIFIER constexpr
-#else
-// The qualifier is `const` instead of `constexpr` that is original to be
-// because the constructor is not `constexpr` function.
-#define CONSTEXPR_QUALIFIER const
-#endif
 
 // Partial specialization of `std::hash<cl::sycl::half>`
 template <> struct hash<half> {
@@ -307,35 +306,43 @@ template <> struct numeric_limits<half> {
 
   static constexpr const float_round_style round_style = round_to_nearest;
 
-  static CONSTEXPR_QUALIFIER half min() noexcept { return SYCL_HLF_MIN; }
+  static __SYCL_CONSTEXPR_ON_DEVICE const half min() noexcept {
+    return SYCL_HLF_MIN;
+  }
 
-  static CONSTEXPR_QUALIFIER half max() noexcept { return SYCL_HLF_MAX; }
+  static __SYCL_CONSTEXPR_ON_DEVICE const half max() noexcept {
+    return SYCL_HLF_MAX;
+  }
 
-  static CONSTEXPR_QUALIFIER half lowest() noexcept { return -SYCL_HLF_MAX; }
+  static __SYCL_CONSTEXPR_ON_DEVICE const half lowest() noexcept {
+    return -SYCL_HLF_MAX;
+  }
 
-  static CONSTEXPR_QUALIFIER half epsilon() noexcept {
+  static __SYCL_CONSTEXPR_ON_DEVICE const half epsilon() noexcept {
     return SYCL_HLF_EPSILON;
   }
 
-  static CONSTEXPR_QUALIFIER half round_error() noexcept { return 0.5F; }
+  static __SYCL_CONSTEXPR_ON_DEVICE const half round_error() noexcept {
+    return 0.5F;
+  }
 
-  static CONSTEXPR_QUALIFIER half infinity() noexcept {
+  static __SYCL_CONSTEXPR_ON_DEVICE const half infinity() noexcept {
     return __builtin_huge_valf();
   }
 
-  static CONSTEXPR_QUALIFIER half quiet_NaN() noexcept {
+  static __SYCL_CONSTEXPR_ON_DEVICE const half quiet_NaN() noexcept {
     return __builtin_nanf("");
   }
 
-  static CONSTEXPR_QUALIFIER half signaling_NaN() noexcept {
+  static __SYCL_CONSTEXPR_ON_DEVICE const half signaling_NaN() noexcept {
     return __builtin_nansf("");
   }
 
-  static CONSTEXPR_QUALIFIER half denorm_min() noexcept { return 5.96046e-08F; }
+  static __SYCL_CONSTEXPR_ON_DEVICE const half denorm_min() noexcept {
+    return 5.96046e-08F;
+  }
 };
 
-#undef CONSTEXPR_QUALIFIER
-
 } // namespace std
 
 inline std::ostream &operator<<(std::ostream &O, half const &rhs) {
@@ -349,3 +356,5 @@ inline std::istream &operator>>(std::istream &I, half &rhs) {
   rhs = ValFloat;
   return I;
 }
+
+#undef __SYCL_CONSTEXPR_ON_DEVICE