Move accessor_impl to sources

Author: romanovvlad

sycl/include/sycl/accessor.hpp

@@ -11,7 +11,6 @@
 #include <CL/__spirv/spirv_types.hpp>
 #include <sycl/atomic.hpp>
 #include <sycl/buffer.hpp>
-#include <sycl/detail/accessor_impl.hpp>
 #include <sycl/detail/cl.h>
 #include <sycl/detail/common.hpp>
 #include <sycl/detail/export.hpp>
@@ -230,6 +229,7 @@ void __SYCL_EXPORT constructorNotification(void *BufferObj, void *AccessorObj,
                                            access::target Target,
                                            access::mode Mode,
                                            const code_location &CodeLoc);
+
 template <typename T>
 using IsPropertyListT = typename std::is_base_of<PropertyListBase, T>;
 
@@ -340,6 +340,159 @@ class accessor_common {
   };
 };
 
+#if __cplusplus >= 201703L
+
+template <typename MayBeTag1, typename MayBeTag2>
+constexpr access::mode deduceAccessMode() {
+  // property_list = {} is not properly detected by deduction guide,
+  // when parameter is passed without curly braces: access(buffer, no_init)
+  // thus simplest approach is to check 2 last arguments for being a tag
+  if constexpr (std::is_same<MayBeTag1,
+                             mode_tag_t<access::mode::read>>::value ||
+                std::is_same<MayBeTag2,
+                             mode_tag_t<access::mode::read>>::value) {
+    return access::mode::read;
+  }
+
+  if constexpr (std::is_same<MayBeTag1,
+                             mode_tag_t<access::mode::write>>::value ||
+                std::is_same<MayBeTag2,
+                             mode_tag_t<access::mode::write>>::value) {
+    return access::mode::write;
+  }
+
+  if constexpr (
+      std::is_same<MayBeTag1,
+                   mode_target_tag_t<access::mode::read,
+                                     access::target::constant_buffer>>::value ||
+      std::is_same<MayBeTag2,
+                   mode_target_tag_t<access::mode::read,
+                                     access::target::constant_buffer>>::value) {
+    return access::mode::read;
+  }
+
+  return access::mode::read_write;
+}
+
+template <typename MayBeTag1, typename MayBeTag2>
+constexpr access::target deduceAccessTarget(access::target defaultTarget) {
+  if constexpr (
+      std::is_same<MayBeTag1,
+                   mode_target_tag_t<access::mode::read,
+                                     access::target::constant_buffer>>::value ||
+      std::is_same<MayBeTag2,
+                   mode_target_tag_t<access::mode::read,
+                                     access::target::constant_buffer>>::value) {
+    return access::target::constant_buffer;
+  }
+
+  return defaultTarget;
+}
+
+#endif
+
+
+
+template <int Dims> class LocalAccessorBaseDevice {
+public:
+  LocalAccessorBaseDevice(sycl::range<Dims> Size)
+      : AccessRange(Size),
+        MemRange(InitializedVal<Dims, range>::template get<0>()) {}
+  // TODO: Actually we need only one field here, but currently compiler requires
+  // all of them.
+  range<Dims> AccessRange;
+  range<Dims> MemRange;
+  id<Dims> Offset;
+
+  bool operator==(const LocalAccessorBaseDevice &Rhs) const {
+    return (AccessRange == Rhs.AccessRange);
+  }
+};
+
+// The class describes a requirement to access a SYCL memory object such as
+// sycl::buffer and sycl::image. For example, each accessor used in a kernel,
+// except one with access target "local", adds such requirement for the command
+// group.
+
+template <int Dims> class AccessorImplDevice {
+public:
+  AccessorImplDevice() = default;
+  AccessorImplDevice(id<Dims> Offset, range<Dims> AccessRange,
+                     range<Dims> MemoryRange)
+      : Offset(Offset), AccessRange(AccessRange), MemRange(MemoryRange) {}
+
+  id<Dims> Offset;
+  range<Dims> AccessRange;
+  range<Dims> MemRange;
+
+  bool operator==(const AccessorImplDevice &Rhs) const {
+    return (Offset == Rhs.Offset && AccessRange == Rhs.AccessRange &&
+            MemRange == Rhs.MemRange);
+  }
+};
+
+class AccessorImplHost;
+
+void __SYCL_EXPORT addHostAccessorAndWait(AccessorImplHost *Req);
+
+class SYCLMemObjI;
+
+using AccessorImplPtr = std::shared_ptr<AccessorImplHost>;
+
+class __SYCL_EXPORT AccessorBaseHost {
+public:
+  AccessorBaseHost(id<3> Offset, range<3> AccessRange, range<3> MemoryRange,
+                   access::mode AccessMode, void *SYCLMemObject, int Dims,
+                   int ElemSize, int OffsetInBytes = 0,
+                   bool IsSubBuffer = false);
+
+protected:
+  id<3> &getOffset();
+  range<3> &getAccessRange();
+  range<3> &getMemoryRange();
+  void *getPtr();
+  unsigned int getElemSize() const;
+
+  const id<3> &getOffset() const;
+  const range<3> &getAccessRange() const;
+  const range<3> &getMemoryRange() const;
+  void *getPtr() const;
+
+  void *getMemoryObject() const;
+
+  template <class Obj>
+  friend decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject);
+
+  template <typename, int, access::mode, access::target, access::placeholder,
+            typename>
+  friend class accessor;
+
+  AccessorImplPtr impl;
+
+private:
+  friend class sycl::ext::intel::esimd::detail::AccessorPrivateProxy;
+};
+
+class LocalAccessorImplHost;
+using LocalAccessorImplPtr = std::shared_ptr<LocalAccessorImplHost>;
+
+class __SYCL_EXPORT LocalAccessorBaseHost {
+public:
+  LocalAccessorBaseHost(sycl::range<3> Size, int Dims, int ElemSize);
+  sycl::range<3> &getSize();
+  const sycl::range<3> &getSize() const;
+  void *getPtr();
+  void *getPtr() const;
+  int getNumOfDims();
+  int getElementSize();
+
+protected:
+  template <class Obj>
+  friend decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject);
+
+  std::shared_ptr<LocalAccessorImplHost> impl;
+};
+
 template <int Dim, typename T> struct IsValidCoordDataT;
 template <typename T> struct IsValidCoordDataT<1, T> {
   constexpr static bool value =
@@ -1663,8 +1816,8 @@ class __SYCL_SPECIAL_CLASS accessor :
         PropertyListT::template areSameCompileTimeProperties<NewPropsT...>(),
         "Compile-time-constant properties must be the same");
 #ifndef __SYCL_DEVICE_ONLY__
-    detail::constructorNotification(impl.get()->MSYCLMemObj, impl.get(),
-                                    AccessTarget, AccessMode, CodeLoc);
+    detail::constructorNotification(getMemoryObject(), impl.get(), AccessTarget,
+                                    AccessMode, CodeLoc);
 #endif
   }