Revert changes in atomic.hpp - the new implementation can't be used there yet

Signed-off-by: Artem Gindinson <artem.gindinson@intel.com>

Author: Artem Gindinson

sycl/include/CL/sycl/atomic.hpp

@@ -121,12 +121,6 @@ extern T __spirv_AtomicISub(std::atomic<T> *Ptr, __spv::Scope::Flag,
   return Ptr->fetch_sub(V, ::cl::sycl::detail::getStdMemoryOrder(MS));
 }
 
-template <typename T>
-extern T __spirv_AtomicFAdd(std::atomic<T> *Ptr, __spv::Scope::Flag,
-                            __spv::MemorySemanticsMask::Flag MS, T V) {
-  return Ptr->fetch_add(V, ::cl::sycl::detail::getStdMemoryOrder(MS));
-}
-
 template <typename T>
 extern T __spirv_AtomicAnd(std::atomic<T> *Ptr, __spv::Scope::Flag,
                            __spv::MemorySemanticsMask::Flag MS, T V) {
@@ -280,36 +274,18 @@ class atomic {
 #endif
   }
 
-  template <typename T2 = T>
-  detail::enable_if_t<std::is_integral<T2>::value, T>
-  fetch_add(T Operand, memory_order Order = memory_order::relaxed) {
+  T fetch_add(T Operand, memory_order Order = memory_order::relaxed) {
+    __SYCL_STATIC_ASSERT_NOT_FLOAT(T);
     return __spirv_AtomicIAdd(
         Ptr, SpirvScope, detail::getSPIRVMemorySemanticsMask(Order), Operand);
   }
 
-  template <typename T2 = T>
-  detail::enable_if_t<std::is_floating_point<T2>::value, T>
-  fetch_add(T Operand, memory_order Order = memory_order::relaxed) {
-    return __spirv_AtomicFAdd(
-        Ptr, SpirvScope, detail::getSPIRVMemorySemanticsMask(Order), Operand);
-  }
-
-  template <typename T2 = T>
-  detail::enable_if_t<std::is_integral<T2>::value, T>
-  fetch_sub(T Operand, memory_order Order = memory_order::relaxed) {
+  T fetch_sub(T Operand, memory_order Order = memory_order::relaxed) {
     __SYCL_STATIC_ASSERT_NOT_FLOAT(T);
     return __spirv_AtomicISub(
         Ptr, SpirvScope, detail::getSPIRVMemorySemanticsMask(Order), Operand);
   }
 
-  template <typename T2 = T>
-  detail::enable_if_t<std::is_floating_point<T2>::value, T>
-  fetch_sub(T Operand, memory_order Order = memory_order::relaxed) {
-    // Negate the float add instruction operand.
-    return __spirv_AtomicFAdd(
-        Ptr, SpirvScope, detail::getSPIRVMemorySemanticsMask(Order), -Operand);
-  }
-
   T fetch_and(T Operand, memory_order Order = memory_order::relaxed) {
     __SYCL_STATIC_ASSERT_NOT_FLOAT(T);
     return __spirv_AtomicAnd(