1 /*
2 //@HEADER
3 // ************************************************************************
4 //
5 //                        Kokkos v. 3.0
6 //       Copyright (2020) National Technology & Engineering
7 //               Solutions of Sandia, LLC (NTESS).
8 //
9 // Under the terms of Contract DE-NA0003525 with NTESS,
10 // the U.S. Government retains certain rights in this software.
11 //
12 // Redistribution and use in source and binary forms, with or without
13 // modification, are permitted provided that the following conditions are
14 // met:
15 //
16 // 1. Redistributions of source code must retain the above copyright
17 // notice, this list of conditions and the following disclaimer.
18 //
19 // 2. Redistributions in binary form must reproduce the above copyright
20 // notice, this list of conditions and the following disclaimer in the
21 // documentation and/or other materials provided with the distribution.
22 //
23 // 3. Neither the name of the Corporation nor the names of the
24 // contributors may be used to endorse or promote products derived from
25 // this software without specific prior written permission.
26 //
27 // THIS SOFTWARE IS PROVIDED BY NTESS "AS IS" AND ANY
28 // EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
29 // IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR
30 // PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL NTESS OR THE
31 // CONTRIBUTORS BE LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL,
32 // EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO,
33 // PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR
34 // PROFITS; OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF
35 // LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING
36 // NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE OF THIS
37 // SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
38 //
39 // Questions? Contact Christian R. Trott (crtrott@sandia.gov)
40 //
41 // ************************************************************************
42 //@HEADER
43 */
44 
45 #ifndef KOKKOS_SYCLSPACE_HPP
46 #define KOKKOS_SYCLSPACE_HPP
47 
48 #include <Kokkos_Core_fwd.hpp>
49 
50 #ifdef KOKKOS_ENABLE_SYCL
51 #include <Kokkos_Concepts.hpp>
52 #include <Kokkos_ScratchSpace.hpp>
53 #include <SYCL/Kokkos_SYCL_Instance.hpp>
54 #include <impl/Kokkos_SharedAlloc.hpp>
55 #include <impl/Kokkos_Tools.hpp>
56 
57 namespace Kokkos {
58 namespace Experimental {
59 
60 class SYCLDeviceUSMSpace {
61  public:
62   using execution_space = SYCL;
63   using memory_space    = SYCLDeviceUSMSpace;
64   using device_type     = Kokkos::Device<execution_space, memory_space>;
65   using size_type       = Impl::SYCLInternal::size_type;
66 
67   SYCLDeviceUSMSpace();
68   explicit SYCLDeviceUSMSpace(sycl::queue queue);
69 
70   void* allocate(const std::size_t arg_alloc_size) const;
71   void* allocate(const char* arg_label, const size_t arg_alloc_size,
72                  const size_t arg_logical_size = 0) const;
73 
74   void deallocate(void* const arg_alloc_ptr,
75                   const std::size_t arg_alloc_size) const;
76   void deallocate(const char* arg_label, void* const arg_alloc_ptr,
77                   const size_t arg_alloc_size,
78                   const size_t arg_logical_size = 0) const;
79 
80  private:
81   template <class, class, class, class>
82   friend class LogicalMemorySpace;
83 
84  public:
name()85   static constexpr const char* name() { return "SYCLDeviceUSM"; };
86 
87  private:
88   sycl::queue m_queue;
89 };
90 
91 class SYCLSharedUSMSpace {
92  public:
93   using execution_space = SYCL;
94   using memory_space    = SYCLSharedUSMSpace;
95   using device_type     = Kokkos::Device<execution_space, memory_space>;
96   using size_type       = Impl::SYCLInternal::size_type;
97 
98   SYCLSharedUSMSpace();
99   explicit SYCLSharedUSMSpace(sycl::queue queue);
100 
101   void* allocate(const std::size_t arg_alloc_size) const;
102   void* allocate(const char* arg_label, const size_t arg_alloc_size,
103                  const size_t arg_logical_size = 0) const;
104 
105   void deallocate(void* const arg_alloc_ptr,
106                   const std::size_t arg_alloc_size) const;
107   void deallocate(const char* arg_label, void* const arg_alloc_ptr,
108                   const size_t arg_alloc_size,
109                   const size_t arg_logical_size = 0) const;
110 
111  private:
112   template <class, class, class, class>
113   friend class LogicalMemorySpace;
114 
115  public:
name()116   static constexpr const char* name() { return "SYCLSharedUSM"; };
117 
118  private:
119   sycl::queue m_queue;
120 };
121 }  // namespace Experimental
122 
123 namespace Impl {
124 static_assert(Kokkos::Impl::MemorySpaceAccess<
125                   Kokkos::Experimental::SYCLDeviceUSMSpace,
126                   Kokkos::Experimental::SYCLDeviceUSMSpace>::assignable,
127               "");
128 
129 static_assert(Kokkos::Impl::MemorySpaceAccess<
130                   Kokkos::Experimental::SYCLSharedUSMSpace,
131                   Kokkos::Experimental::SYCLSharedUSMSpace>::assignable,
132               "");
133 
134 template <>
135 struct MemorySpaceAccess<Kokkos::HostSpace,
136                          Kokkos::Experimental::SYCLDeviceUSMSpace> {
137   enum : bool { assignable = false };
138   enum : bool { accessible = false };
139   enum : bool { deepcopy = true };
140 };
141 
142 template <>
143 struct MemorySpaceAccess<Kokkos::HostSpace,
144                          Kokkos::Experimental::SYCLSharedUSMSpace> {
145   // HostSpace::execution_space != SYCLSharedUSMSpace::execution_space
146   enum : bool { assignable = false };
147   enum : bool { accessible = true };
148   enum : bool { deepcopy = true };
149 };
150 
151 template <>
152 struct MemorySpaceAccess<Kokkos::Experimental::SYCLDeviceUSMSpace,
153                          Kokkos::HostSpace> {
154   enum : bool { assignable = false };
155   enum : bool { accessible = false };
156   enum : bool { deepcopy = true };
157 };
158 
159 template <>
160 struct MemorySpaceAccess<Kokkos::Experimental::SYCLDeviceUSMSpace,
161                          Kokkos::Experimental::SYCLSharedUSMSpace> {
162   // SYCLDeviceUSMSpace::execution_space == SYCLSharedUSMSpace::execution_space
163   enum : bool { assignable = true };
164   enum : bool { accessible = true };
165   enum : bool { deepcopy = true };
166 };
167 
168 //----------------------------------------
169 // SYCLSharedUSMSpace::execution_space == SYCL
170 // SYCLSharedUSMSpace accessible to both SYCL and Host
171 
172 template <>
173 struct MemorySpaceAccess<Kokkos::Experimental::SYCLSharedUSMSpace,
174                          Kokkos::HostSpace> {
175   enum : bool { assignable = false };
176   enum : bool { accessible = false };  // SYCL cannot access HostSpace
177   enum : bool { deepcopy = true };
178 };
179 
180 template <>
181 struct MemorySpaceAccess<Kokkos::Experimental::SYCLSharedUSMSpace,
182                          Kokkos::Experimental::SYCLDeviceUSMSpace> {
183   // SYCLSharedUSMSpace::execution_space == SYCLDeviceUSMSpace::execution_space
184   // Can access SYCLSharedUSMSpace from Host but cannot access
185   // SYCLDeviceUSMSpace from Host
186   enum : bool { assignable = false };
187 
188   // SYCLSharedUSMSpace::execution_space can access SYCLDeviceUSMSpace
189   enum : bool { accessible = true };
190   enum : bool { deepcopy = true };
191 };
192 
193 template <>
194 struct MemorySpaceAccess<
195     Kokkos::Experimental::SYCLDeviceUSMSpace,
196     Kokkos::ScratchMemorySpace<Kokkos::Experimental::SYCL>> {
197   enum : bool { assignable = false };
198   enum : bool { accessible = true };
199   enum : bool { deepcopy = false };
200 };
201 
202 template <>
203 struct MemorySpaceAccess<
204     Kokkos::Experimental::SYCLSharedUSMSpace,
205     Kokkos::ScratchMemorySpace<Kokkos::Experimental::SYCL>> {
206   enum : bool { assignable = false };
207   enum : bool { accessible = true };
208   enum : bool { deepcopy = false };
209 };
210 
211 }  // namespace Impl
212 
213 namespace Impl {
214 
215 template <>
216 class SharedAllocationRecord<Kokkos::Experimental::SYCLDeviceUSMSpace, void>
217     : public HostInaccessibleSharedAllocationRecordCommon<
218           Kokkos::Experimental::SYCLDeviceUSMSpace> {
219  private:
220   friend class SharedAllocationRecordCommon<
221       Kokkos::Experimental::SYCLDeviceUSMSpace>;
222   friend class HostInaccessibleSharedAllocationRecordCommon<
223       Kokkos::Experimental::SYCLDeviceUSMSpace>;
224   using base_t = HostInaccessibleSharedAllocationRecordCommon<
225       Kokkos::Experimental::SYCLDeviceUSMSpace>;
226   using RecordBase = SharedAllocationRecord<void, void>;
227 
228   SharedAllocationRecord(const SharedAllocationRecord&) = delete;
229   SharedAllocationRecord(SharedAllocationRecord&&)      = delete;
230   SharedAllocationRecord& operator=(const SharedAllocationRecord&) = delete;
231   SharedAllocationRecord& operator=(SharedAllocationRecord&&) = delete;
232 
233 #ifdef KOKKOS_ENABLE_DEBUG
234   static RecordBase s_root_record;
235 #endif
236 
237   const Kokkos::Experimental::SYCLDeviceUSMSpace m_space;
238 
239  protected:
240   ~SharedAllocationRecord();
241 
242   SharedAllocationRecord(
243       const Kokkos::Experimental::SYCLDeviceUSMSpace& arg_space,
244       const std::string& arg_label, const size_t arg_alloc_size,
245       const RecordBase::function_type arg_dealloc = &base_t::deallocate);
246 };
247 
248 template <>
249 class SharedAllocationRecord<Kokkos::Experimental::SYCLSharedUSMSpace, void>
250     : public SharedAllocationRecordCommon<
251           Kokkos::Experimental::SYCLSharedUSMSpace> {
252  private:
253   friend class SharedAllocationRecordCommon<
254       Kokkos::Experimental::SYCLSharedUSMSpace>;
255   using base_t =
256       SharedAllocationRecordCommon<Kokkos::Experimental::SYCLSharedUSMSpace>;
257   using RecordBase = SharedAllocationRecord<void, void>;
258 
259   SharedAllocationRecord(const SharedAllocationRecord&) = delete;
260   SharedAllocationRecord(SharedAllocationRecord&&)      = delete;
261   SharedAllocationRecord& operator=(const SharedAllocationRecord&) = delete;
262   SharedAllocationRecord& operator=(SharedAllocationRecord&&) = delete;
263 
264   static RecordBase s_root_record;
265 
266   const Kokkos::Experimental::SYCLSharedUSMSpace m_space;
267 
268  protected:
269   ~SharedAllocationRecord();
270 
271   SharedAllocationRecord() = default;
272 
273   SharedAllocationRecord(
274       const Kokkos::Experimental::SYCLSharedUSMSpace& arg_space,
275       const std::string& arg_label, const size_t arg_alloc_size,
276       const RecordBase::function_type arg_dealloc = &base_t::deallocate);
277 };
278 
279 }  // namespace Impl
280 
281 }  // namespace Kokkos
282 
283 #endif
284 #endif
285