Kokkos Core Kernels Package Version of the Day
Loading...
Searching...
No Matches
Kokkos_SYCL.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_SYCL_HPP
23#define KOKKOS_SYCL_HPP
24
25#include <Kokkos_Macros.hpp>
26
27#ifdef KOKKOS_ENABLE_SYCL
28// FIXME_SYCL
29#if __has_include(<sycl/sycl.hpp>)
30#include <sycl/sycl.hpp>
31#else
32#include <CL/sycl.hpp>
33#endif
34#include <Kokkos_SYCL_Space.hpp>
35#include <Kokkos_Layout.hpp>
36#include <Kokkos_ScratchSpace.hpp>
37#include <impl/Kokkos_Profiling_Interface.hpp>
38#include <impl/Kokkos_HostSharedPtr.hpp>
39#include <impl/Kokkos_InitializationSettings.hpp>
40
41namespace Kokkos {
42namespace Experimental {
43namespace Impl {
44class SYCLInternal;
45}
46
49class SYCL {
50 public:
51 //------------------------------------
53
54
56 using execution_space = SYCL;
57 using memory_space = SYCLDeviceUSMSpace;
59
60 using array_layout = LayoutLeft;
61 using size_type = memory_space::size_type;
62
63 using scratch_memory_space = ScratchMemorySpace<SYCL>;
64
65 SYCL();
66 explicit SYCL(const sycl::queue&);
67
68 uint32_t impl_instance_id() const noexcept {
69 return m_space_instance->impl_get_instance_id();
70 }
71
72 sycl::queue& sycl_queue() const noexcept {
73 return *m_space_instance->m_queue;
74 }
75
77 //------------------------------------
79
80
81 KOKKOS_INLINE_FUNCTION static int in_parallel() {
82#if defined(__SYCL_DEVICE_ONLY__)
83 return true;
84#else
85 return false;
86#endif
87 }
88
90 static bool sleep();
91
93 static bool wake();
94
96 static void impl_static_fence(const std::string& name);
97
98 void fence(
99 const std::string& name =
100 "Kokkos::Experimental::SYCL::fence: Unnamed Instance Fence") const;
101
103 void print_configuration(std::ostream& os, bool verbose = false) const;
104
106 static void impl_finalize();
107
108 static void impl_initialize(InitializationSettings const&);
109
110 static bool impl_is_initialized();
111
112#ifdef KOKKOS_ENABLE_DEPRECATED_CODE_4
113 static int concurrency();
114#else
115 int concurrency() const;
116#endif
117
118 static const char* name();
119
120 inline Impl::SYCLInternal* impl_internal_space_instance() const {
121 return m_space_instance.get();
122 }
123
124 private:
125 static std::ostream& impl_sycl_info(std::ostream& os,
126 const sycl::device& device);
127
128 friend bool operator==(SYCL const& lhs, SYCL const& rhs) {
129 return lhs.impl_internal_space_instance() ==
130 rhs.impl_internal_space_instance();
131 }
132 friend bool operator!=(SYCL const& lhs, SYCL const& rhs) {
133 return !(lhs == rhs);
134 }
135 Kokkos::Impl::HostSharedPtr<Impl::SYCLInternal> m_space_instance;
136};
137
138} // namespace Experimental
139
140namespace Tools {
141namespace Experimental {
142template <>
143struct DeviceTypeTraits<Kokkos::Experimental::SYCL> {
145 static constexpr DeviceType id = DeviceType::SYCL;
146 static int device_id(const Kokkos::Experimental::SYCL& exec) {
147 return exec.impl_internal_space_instance()->m_syclDev;
148 }
149};
150} // namespace Experimental
151} // namespace Tools
152
153namespace Experimental {
154template <class... Args>
155std::vector<SYCL> partition_space(const SYCL& sycl_space, Args...) {
156 static_assert(
157 (... && std::is_arithmetic_v<Args>),
158 "Kokkos Error: partitioning arguments must be integers or floats");
159
160 sycl::context context = sycl_space.sycl_queue().get_context();
161 sycl::device device =
162 sycl_space.impl_internal_space_instance()->m_queue->get_device();
163 std::vector<SYCL> instances;
164 instances.reserve(sizeof...(Args));
165 for (unsigned int i = 0; i < sizeof...(Args); ++i)
166 instances.emplace_back(sycl::queue(context, device));
167 return instances;
168}
169
170template <class T>
171std::vector<SYCL> partition_space(const SYCL& sycl_space,
172 std::vector<T>& weights) {
173 static_assert(
174 std::is_arithmetic<T>::value,
175 "Kokkos Error: partitioning arguments must be integers or floats");
176
177 sycl::context context = sycl_space.sycl_queue().get_context();
178 sycl::device device =
179 sycl_space.impl_internal_space_instance()->m_queue->get_device();
180 std::vector<SYCL> instances;
181 instances.reserve(weights.size());
182 for (unsigned int i = 0; i < weights.size(); ++i)
183 instances.emplace_back(sycl::queue(context, device));
184 return instances;
185}
186} // namespace Experimental
187
188} // namespace Kokkos
189
190#endif
191#endif
Declaration of various MemoryLayout options.
A thread safe view to a bitset.