Kokkos Core Kernels Package Version of the Day
Loading...
Searching...
No Matches
Kokkos_SYCL_Space.hpp
1//@HEADER
2// ************************************************************************
3//
4// Kokkos v. 4.0
5// Copyright (2022) National Technology & Engineering
6// Solutions of Sandia, LLC (NTESS).
7//
8// Under the terms of Contract DE-NA0003525 with NTESS,
9// the U.S. Government retains certain rights in this software.
10//
11// Part of Kokkos, under the Apache License v2.0 with LLVM Exceptions.
12// See https://kokkos.org/LICENSE for license information.
13// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
14//
15//@HEADER
16
17#ifndef KOKKOS_IMPL_PUBLIC_INCLUDE
18#include <Kokkos_Macros.hpp>
19static_assert(false,
20 "Including non-public Kokkos header files is not allowed.");
21#endif
22#ifndef KOKKOS_SYCLSPACE_HPP
23#define KOKKOS_SYCLSPACE_HPP
24
25#include <Kokkos_Core_fwd.hpp>
26
27#ifdef KOKKOS_ENABLE_SYCL
28#include <Kokkos_Concepts.hpp>
29#include <Kokkos_HostSpace.hpp>
30#include <Kokkos_ScratchSpace.hpp>
31#include <SYCL/Kokkos_SYCL_Instance.hpp>
32#include <impl/Kokkos_SharedAlloc.hpp>
33#include <impl/Kokkos_Tools.hpp>
34
35namespace Kokkos {
36
37namespace Impl {
38template <typename T>
39struct is_sycl_type_space : public std::false_type {};
40} // namespace Impl
41
42namespace Experimental {
43
44class SYCLDeviceUSMSpace {
45 public:
46 using execution_space = SYCL;
47 using memory_space = SYCLDeviceUSMSpace;
49 using size_type = Impl::SYCLInternal::size_type;
50
51 SYCLDeviceUSMSpace();
52 explicit SYCLDeviceUSMSpace(sycl::queue queue);
53
54 void* allocate(const SYCL& exec_space,
55 const std::size_t arg_alloc_size) const;
56 void* allocate(const SYCL& exec_space, const char* arg_label,
57 const size_t arg_alloc_size,
58 const size_t arg_logical_size = 0) const;
59 void* allocate(const std::size_t arg_alloc_size) const;
60 void* allocate(const char* arg_label, const size_t arg_alloc_size,
61 const size_t arg_logical_size = 0) const;
62
63 void deallocate(void* const arg_alloc_ptr,
64 const std::size_t arg_alloc_size) const;
65 void deallocate(const char* arg_label, void* const arg_alloc_ptr,
66 const size_t arg_alloc_size,
67 const size_t arg_logical_size = 0) const;
68
69 private:
70 template <class, class, class, class>
71 friend class LogicalMemorySpace;
72
73 public:
74 static constexpr const char* name() { return "SYCLDeviceUSM"; };
75
76 private:
77 sycl::queue m_queue;
78};
79
80class SYCLSharedUSMSpace {
81 public:
82 using execution_space = SYCL;
83 using memory_space = SYCLSharedUSMSpace;
85 using size_type = Impl::SYCLInternal::size_type;
86
87 SYCLSharedUSMSpace();
88 explicit SYCLSharedUSMSpace(sycl::queue queue);
89
90 void* allocate(const SYCL& exec_space,
91 const std::size_t arg_alloc_size) const;
92 void* allocate(const SYCL& exec_space, const char* arg_label,
93 const size_t arg_alloc_size,
94 const size_t arg_logical_size = 0) const;
95 void* allocate(const std::size_t arg_alloc_size) const;
96 void* allocate(const char* arg_label, const size_t arg_alloc_size,
97 const size_t arg_logical_size = 0) const;
98
99 void deallocate(void* const arg_alloc_ptr,
100 const std::size_t arg_alloc_size) const;
101 void deallocate(const char* arg_label, void* const arg_alloc_ptr,
102 const size_t arg_alloc_size,
103 const size_t arg_logical_size = 0) const;
104
105 private:
106 template <class, class, class, class>
107 friend class LogicalMemorySpace;
108
109 public:
110 static constexpr const char* name() { return "SYCLSharedUSM"; };
111
112 private:
113 sycl::queue m_queue;
114};
115
116class SYCLHostUSMSpace {
117 public:
118 using execution_space = HostSpace::execution_space;
119 using memory_space = SYCLHostUSMSpace;
121 using size_type = Impl::SYCLInternal::size_type;
122
123 SYCLHostUSMSpace();
124 explicit SYCLHostUSMSpace(sycl::queue queue);
125
126 void* allocate(const SYCL& exec_space,
127 const std::size_t arg_alloc_size) const;
128 void* allocate(const SYCL& exec_space, const char* arg_label,
129 const size_t arg_alloc_size,
130 const size_t arg_logical_size = 0) const;
131 void* allocate(const std::size_t arg_alloc_size) const;
132 void* allocate(const char* arg_label, const size_t arg_alloc_size,
133 const size_t arg_logical_size = 0) const;
134
135 void deallocate(void* const arg_alloc_ptr,
136 const std::size_t arg_alloc_size) const;
137 void deallocate(const char* arg_label, void* const arg_alloc_ptr,
138 const size_t arg_alloc_size,
139 const size_t arg_logical_size = 0) const;
140
141 private:
142 template <class, class, class, class>
143 friend class LogicalMemorySpace;
144
145 public:
146 static constexpr const char* name() { return "SYCLHostUSM"; };
147
148 private:
149 sycl::queue m_queue;
150};
151
152} // namespace Experimental
153
154namespace Impl {
155
156template <>
157struct is_sycl_type_space<Kokkos::Experimental::SYCLDeviceUSMSpace>
158 : public std::true_type {};
159
160template <>
161struct is_sycl_type_space<Kokkos::Experimental::SYCLSharedUSMSpace>
162 : public std::true_type {};
163
164template <>
165struct is_sycl_type_space<Kokkos::Experimental::SYCLHostUSMSpace>
166 : public std::true_type {};
167
169 Kokkos::Experimental::SYCLDeviceUSMSpace,
170 Kokkos::Experimental::SYCLDeviceUSMSpace>::assignable,
171 "");
172
174 Kokkos::Experimental::SYCLSharedUSMSpace,
175 Kokkos::Experimental::SYCLSharedUSMSpace>::assignable,
176 "");
177
179 Kokkos::Experimental::SYCLDeviceUSMSpace,
180 Kokkos::Experimental::SYCLDeviceUSMSpace>::assignable,
181 "");
182
183template <>
184struct MemorySpaceAccess<Kokkos::HostSpace,
185 Kokkos::Experimental::SYCLDeviceUSMSpace> {
186 enum : bool { assignable = false };
187 enum : bool { accessible = false };
188 enum : bool { deepcopy = true };
189};
190
191template <>
192struct MemorySpaceAccess<Kokkos::HostSpace,
193 Kokkos::Experimental::SYCLSharedUSMSpace> {
194 // HostSpace::execution_space != SYCLSharedUSMSpace::execution_space
195 enum : bool { assignable = false };
196 enum : bool { accessible = true };
197 enum : bool { deepcopy = true };
198};
199
200template <>
201struct MemorySpaceAccess<Kokkos::HostSpace,
202 Kokkos::Experimental::SYCLHostUSMSpace> {
203 // HostSpace::execution_space ==
204 // Experimental::SYCLHostUSMSpace::execution_space
205 enum : bool { assignable = true };
206 enum : bool { accessible = true };
207 enum : bool { deepcopy = true };
208};
209
210template <>
211struct MemorySpaceAccess<Kokkos::Experimental::SYCLDeviceUSMSpace,
213 enum : bool { assignable = false };
214 enum : bool { accessible = false };
215 enum : bool { deepcopy = true };
216};
217
218template <>
219struct MemorySpaceAccess<Kokkos::Experimental::SYCLDeviceUSMSpace,
220 Kokkos::Experimental::SYCLSharedUSMSpace> {
221 // SYCLDeviceUSMSpace::execution_space == SYCLSharedUSMSpace::execution_space
222 enum : bool { assignable = true };
223 enum : bool { accessible = true };
224 enum : bool { deepcopy = true };
225};
226
227template <>
228struct MemorySpaceAccess<Kokkos::Experimental::SYCLDeviceUSMSpace,
229 Kokkos::Experimental::SYCLHostUSMSpace> {
230 // Experimental::SYCLDeviceUSMSpace::execution_space !=
231 // Experimental::SYCLHostUSMSpace::execution_space
232 enum : bool { assignable = false };
233 enum : bool {
234 accessible = true
235 }; // Experimental::SYCLDeviceUSMSpace::execution_space
236 enum : bool { deepcopy = true };
237};
238
239//----------------------------------------
240// SYCLSharedUSMSpace::execution_space == SYCL
241// SYCLSharedUSMSpace accessible to both SYCL and Host
242
243template <>
244struct MemorySpaceAccess<Kokkos::Experimental::SYCLSharedUSMSpace,
246 enum : bool { assignable = false };
247 enum : bool { accessible = false }; // SYCL cannot access HostSpace
248 enum : bool { deepcopy = true };
249};
250
251template <>
252struct MemorySpaceAccess<Kokkos::Experimental::SYCLSharedUSMSpace,
253 Kokkos::Experimental::SYCLDeviceUSMSpace> {
254 // SYCLSharedUSMSpace::execution_space == SYCLDeviceUSMSpace::execution_space
255 // Can access SYCLSharedUSMSpace from Host but cannot access
256 // SYCLDeviceUSMSpace from Host
257 enum : bool { assignable = false };
258
259 // SYCLSharedUSMSpace::execution_space can access SYCLDeviceUSMSpace
260 enum : bool { accessible = true };
261 enum : bool { deepcopy = true };
262};
263
264template <>
265struct MemorySpaceAccess<Kokkos::Experimental::SYCLSharedUSMSpace,
266 Kokkos::Experimental::SYCLHostUSMSpace> {
267 // Experimental::SYCLSharedUSMSpace::execution_space !=
268 // Experimental::SYCLHostUSMSpace::execution_space
269 enum : bool { assignable = false };
270 enum : bool {
271 accessible = true
272 }; // Experimental::SYCLSharedUSMSpace::execution_space
273 enum : bool { deepcopy = true };
274};
275
276template <>
277struct MemorySpaceAccess<Kokkos::Experimental::SYCLHostUSMSpace,
279 enum : bool { assignable = false }; // Cannot access from SYCL
280 enum : bool {
281 accessible = true
282 }; // Experimental::SYCLHostUSMSpace::execution_space
283 enum : bool { deepcopy = true };
284};
285
286template <>
287struct MemorySpaceAccess<Kokkos::Experimental::SYCLHostUSMSpace,
288 Kokkos::Experimental::SYCLDeviceUSMSpace> {
289 enum : bool { assignable = false }; // Cannot access from Host
290 enum : bool { accessible = false };
291 enum : bool { deepcopy = true };
292};
293
294template <>
295struct MemorySpaceAccess<Kokkos::Experimental::SYCLHostUSMSpace,
296 Kokkos::Experimental::SYCLSharedUSMSpace> {
297 enum : bool { assignable = false }; // different execution_space
298 enum : bool { accessible = true }; // same accessibility
299 enum : bool { deepcopy = true };
300};
301
302template <>
303struct MemorySpaceAccess<
304 Kokkos::Experimental::SYCLDeviceUSMSpace,
305 Kokkos::ScratchMemorySpace<Kokkos::Experimental::SYCL>> {
306 enum : bool { assignable = false };
307 enum : bool { accessible = true };
308 enum : bool { deepcopy = false };
309};
310
311} // namespace Impl
312
313namespace Impl {
314
315template <>
316class SharedAllocationRecord<Kokkos::Experimental::SYCLDeviceUSMSpace, void>
317 : public HostInaccessibleSharedAllocationRecordCommon<
318 Kokkos::Experimental::SYCLDeviceUSMSpace> {
319 private:
320 friend class SharedAllocationRecordCommon<
321 Kokkos::Experimental::SYCLDeviceUSMSpace>;
322 friend class HostInaccessibleSharedAllocationRecordCommon<
323 Kokkos::Experimental::SYCLDeviceUSMSpace>;
324 using base_t = HostInaccessibleSharedAllocationRecordCommon<
325 Kokkos::Experimental::SYCLDeviceUSMSpace>;
326 using RecordBase = SharedAllocationRecord<void, void>;
327
328 SharedAllocationRecord(const SharedAllocationRecord&) = delete;
329 SharedAllocationRecord(SharedAllocationRecord&&) = delete;
330 SharedAllocationRecord& operator=(const SharedAllocationRecord&) = delete;
331 SharedAllocationRecord& operator=(SharedAllocationRecord&&) = delete;
332
333#ifdef KOKKOS_ENABLE_DEBUG
334 static RecordBase s_root_record;
335#endif
336
337 const Kokkos::Experimental::SYCLDeviceUSMSpace m_space;
338
339 protected:
340 ~SharedAllocationRecord();
341
342 template <typename ExecutionSpace>
343 SharedAllocationRecord(
344 const ExecutionSpace& /*exec_space*/,
345 const Kokkos::Experimental::SYCLDeviceUSMSpace& arg_space,
346 const std::string& arg_label, const size_t arg_alloc_size,
347 const RecordBase::function_type arg_dealloc = &base_t::deallocate)
348 : SharedAllocationRecord(arg_space, arg_label, arg_alloc_size,
349 arg_dealloc) {}
350
351 SharedAllocationRecord(
352 const Kokkos::Experimental::SYCL& exec_space,
353 const Kokkos::Experimental::SYCLDeviceUSMSpace& arg_space,
354 const std::string& arg_label, const size_t arg_alloc_size,
355 const RecordBase::function_type arg_dealloc = &base_t::deallocate);
356
357 SharedAllocationRecord(
358 const Kokkos::Experimental::SYCLDeviceUSMSpace& arg_space,
359 const std::string& arg_label, const size_t arg_alloc_size,
360 const RecordBase::function_type arg_dealloc = &base_t::deallocate);
361};
362
363template <>
364class SharedAllocationRecord<Kokkos::Experimental::SYCLSharedUSMSpace, void>
365 : public SharedAllocationRecordCommon<
366 Kokkos::Experimental::SYCLSharedUSMSpace> {
367 private:
368 friend class SharedAllocationRecordCommon<
369 Kokkos::Experimental::SYCLSharedUSMSpace>;
370 using base_t =
371 SharedAllocationRecordCommon<Kokkos::Experimental::SYCLSharedUSMSpace>;
372 using RecordBase = SharedAllocationRecord<void, void>;
373
374 SharedAllocationRecord(const SharedAllocationRecord&) = delete;
375 SharedAllocationRecord(SharedAllocationRecord&&) = delete;
376 SharedAllocationRecord& operator=(const SharedAllocationRecord&) = delete;
377 SharedAllocationRecord& operator=(SharedAllocationRecord&&) = delete;
378
379 static RecordBase s_root_record;
380
381 const Kokkos::Experimental::SYCLSharedUSMSpace m_space;
382
383 protected:
384 ~SharedAllocationRecord();
385
386 SharedAllocationRecord() = default;
387
388 template <typename ExecutionSpace>
389 SharedAllocationRecord(
390 const ExecutionSpace& /*exec_space*/,
391 const Kokkos::Experimental::SYCLSharedUSMSpace& arg_space,
392 const std::string& arg_label, const size_t arg_alloc_size,
393 const RecordBase::function_type arg_dealloc = &base_t::deallocate)
394 : SharedAllocationRecord(arg_space, arg_label, arg_alloc_size,
395 arg_dealloc) {}
396
397 SharedAllocationRecord(
398 const Kokkos::Experimental::SYCL& exec_space,
399 const Kokkos::Experimental::SYCLSharedUSMSpace& arg_space,
400 const std::string& arg_label, const size_t arg_alloc_size,
401 const RecordBase::function_type arg_dealloc = &base_t::deallocate);
402
403 SharedAllocationRecord(
404 const Kokkos::Experimental::SYCLSharedUSMSpace& arg_space,
405 const std::string& arg_label, const size_t arg_alloc_size,
406 const RecordBase::function_type arg_dealloc = &base_t::deallocate);
407};
408
409template <>
410class SharedAllocationRecord<Kokkos::Experimental::SYCLHostUSMSpace, void>
411 : public SharedAllocationRecordCommon<
412 Kokkos::Experimental::SYCLHostUSMSpace> {
413 private:
414 friend class SharedAllocationRecordCommon<
415 Kokkos::Experimental::SYCLHostUSMSpace>;
416 using base_t =
417 SharedAllocationRecordCommon<Kokkos::Experimental::SYCLHostUSMSpace>;
418 using RecordBase = SharedAllocationRecord<void, void>;
419
420 SharedAllocationRecord(const SharedAllocationRecord&) = delete;
421 SharedAllocationRecord(SharedAllocationRecord&&) = delete;
422 SharedAllocationRecord& operator=(const SharedAllocationRecord&) = delete;
423 SharedAllocationRecord& operator=(SharedAllocationRecord&&) = delete;
424
425 static RecordBase s_root_record;
426
427 const Kokkos::Experimental::SYCLHostUSMSpace m_space;
428
429 protected:
430 ~SharedAllocationRecord();
431
432 SharedAllocationRecord() = default;
433
434 template <typename ExecutionSpace>
435 SharedAllocationRecord(
436 const ExecutionSpace& /*exec_space*/,
437 const Kokkos::Experimental::SYCLHostUSMSpace& arg_space,
438 const std::string& arg_label, const size_t arg_alloc_size,
439 const RecordBase::function_type arg_dealloc = &base_t::deallocate)
440 : SharedAllocationRecord(arg_space, arg_label, arg_alloc_size,
441 arg_dealloc) {}
442
443 SharedAllocationRecord(
444 const Kokkos::Experimental::SYCL& exec_space,
445 const Kokkos::Experimental::SYCLHostUSMSpace& arg_space,
446 const std::string& arg_label, const size_t arg_alloc_size,
447 const RecordBase::function_type arg_dealloc = &base_t::deallocate);
448
449 SharedAllocationRecord(
450 const Kokkos::Experimental::SYCLHostUSMSpace& arg_space,
451 const std::string& arg_label, const size_t arg_alloc_size,
452 const RecordBase::function_type arg_dealloc = &base_t::deallocate);
453};
454
455} // namespace Impl
456
457} // namespace Kokkos
458
459#endif
460#endif
A thread safe view to a bitset.
Memory management for host memory.
Scratch memory space associated with an execution space.
Access relationship between DstMemorySpace and SrcMemorySpace.