[SYCL] Add floor and ceil math built-in's

Signed-off-by: Voronov, Alexey <[email protected]>
Signed-off-by: Vladimir Lazarev <[email protected]>

Author: alexeyvoronov-intel

sycl/include/CL/sycl/math.hpp

@@ -175,11 +175,21 @@ GEN_FUNC_OF_THREE_ARG(mad, double, double, double, double)
 // genfloatf exp (genfloatf x)
 GEN_FUNC_OF_ONE_ARG(native_exp, float, float)
 
-// genfloatf fabs (genfloatf x)
+// genfloat fabs (genfloatf x)
 GEN_FUNC_OF_ONE_ARG(fabs, float, float)
 GEN_FUNC_OF_ONE_ARG(fabs, double, double)
 // GEN_FUNC_OF_ONE_ARG(fabs, half, half)
 
+// genfloat floor (genfloat x)
+GEN_FUNC_OF_ONE_ARG(floor, float, float)
+GEN_FUNC_OF_ONE_ARG(floor, double, double)
+// GEN_FUNC_OF_ONE_ARG(floor, half, half)
+
+// genfloat ceil (genfloat x)
+GEN_FUNC_OF_ONE_ARG(ceil, float, float)
+GEN_FUNC_OF_ONE_ARG(ceil, double, double)
+// GEN_FUNC_OF_ONE_ARG(ceil, half, half)
+
 /* --------------- 4.13.4 Integer functions. Device version -----------------*/
 // geninteger max (geninteger x, geninteger y)
 GEN_FUNC_OF_TWO_ARG(max, char, char, char)
@@ -293,6 +303,12 @@ template <typename T> T sqrt(T x) {
 template <typename T> T fabs(T x) {
   return __sycl_std::fabs(x);
 }
+template <typename T> T floor(T x) {
+  return __sycl_std::floor(x);
+}
+template <typename T> T ceil(T x) {
+  return __sycl_std::ceil(x);
+}
 namespace native {
 template <typename T> T exp(T x) {
 #ifdef __SYCL_DEVICE_ONLY__

sycl/test/math/math.cpp

@@ -51,14 +51,15 @@ int main() {
         resultAccessor[wiID.get(0)] += cl::sycl::native::exp(2.f);
         resultAccessor[wiID.get(0)] += cl::sycl::fabs(-2.f);
         resultAccessor[wiID.get(0)] += cl::sycl::fabs(1.0);
+        resultAccessor[wiID.get(0)] += cl::sycl::floor(-3.4);
+        resultAccessor[wiID.get(0)] += cl::sycl::ceil(2.4);
       });
     });
   }
-
   for (size_t i = 0; i < 10; ++i) {
-    /* Result of addition of 2 + 1 + 7.389... + 2 + 1*/
-    assert(result[i] > 13 && result[i] < 14 &&
-           "Expected result[i] > 13 &&  result[i] < 14");
+    /* Result of addition of 2 + 1 + 7.389... + 2 + 1 -4 + 3*/
+    assert(result[i] > 12 && result[i] < 13 &&
+           "Expected result[i] > 12 &&  result[i] < 13");
   }
 
   return 0;