DPC++ Runtime
Runtime libraries for oneAPI DPC++
graph.hpp
Go to the documentation of this file.
1 //==--------- graph.hpp --- SYCL graph extension ---------------------------==//
2 //
3 // Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4 // See https://llvm.org/LICENSE.txt for license information.
5 // SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6 //
7 //===----------------------------------------------------------------------===//
8 
9 #pragma once
10 
11 #include <sycl/accessor.hpp> // for detail::AccessorBaseHost
12 #include <sycl/context.hpp> // for context
13 #include <sycl/detail/export.hpp> // for __SYCL_EXPORT
14 #include <sycl/detail/kernel_desc.hpp> // for kernel_param_kind_t
15 #include <sycl/detail/property_helper.hpp> // for DataLessPropKind, PropWith...
16 #include <sycl/device.hpp> // for device
17 #include <sycl/nd_range.hpp> // for range, nd_range
18 #include <sycl/properties/property_traits.hpp> // for is_property, is_property_of
19 #include <sycl/property_list.hpp> // for property_list
20 
21 #include <functional> // for function
22 #include <memory> // for shared_ptr
23 #include <type_traits> // for true_type
24 #include <vector> // for vector
25 
26 namespace sycl {
27 inline namespace _V1 {
28 
29 class handler;
30 class queue;
31 class device;
32 namespace ext {
33 namespace oneapi {
34 namespace experimental {
35 
37 enum class graph_state {
38  modifiable,
39  executable,
40 };
41 
42 // Forward declare Graph class
43 template <graph_state State> class command_graph;
44 
45 namespace detail {
46 // List of sycl features and extensions which are not supported by graphs. Used
47 // for throwing errors when these features are used with graphs.
49  sycl_reductions = 0,
57 };
58 
59 inline const char *
61  using UGF = UnsupportedGraphFeatures;
62  switch (Feature) {
63  case UGF::sycl_reductions:
64  return "Reductions";
65  case UGF::sycl_specialization_constants:
66  return "Specialization Constants";
67  case UGF::sycl_kernel_bundle:
68  return "Kernel Bundles";
69  case UGF::sycl_ext_oneapi_kernel_properties:
70  return "sycl_ext_oneapi_kernel_properties";
71  case UGF::sycl_ext_oneapi_enqueue_barrier:
72  return "sycl_ext_oneapi_enqueue_barrier";
73  case UGF::sycl_ext_oneapi_memcpy2d:
74  return "sycl_ext_oneapi_memcpy2d";
75  case UGF::sycl_ext_oneapi_device_global:
76  return "sycl_ext_oneapi_device_global";
77  case UGF::sycl_ext_oneapi_bindless_images:
78  return "sycl_ext_oneapi_bindless_images";
79  }
80 
81  assert(false && "Unhandled graphs feature");
82  return {};
83 }
84 
85 class node_impl;
86 class graph_impl;
87 class exec_graph_impl;
88 class dynamic_parameter_impl;
89 } // namespace detail
90 
91 enum class node_type {
92  empty = 0,
93  subgraph = 1,
94  kernel = 2,
95  memcpy = 3,
96  memset = 4,
97  memfill = 5,
98  prefetch = 6,
99  memadvise = 7,
100  ext_oneapi_barrier = 8,
101  host_task = 9
102 };
103 
105 class __SYCL_EXPORT node {
106 public:
107  node() = delete;
108 
110  node_type get_type() const;
111 
113  std::vector<node> get_predecessors() const;
114 
116  std::vector<node> get_successors() const;
117 
120  static node get_node_from_event(event nodeEvent);
121 
123  template <int Dimensions>
124  void update_nd_range(nd_range<Dimensions> executionRange);
125 
127  template <int Dimensions> void update_range(range<Dimensions> executionRange);
128 
129 private:
130  node(const std::shared_ptr<detail::node_impl> &Impl) : impl(Impl) {}
131 
132  template <class Obj>
133  friend decltype(Obj::impl)
134  sycl::detail::getSyclObjImpl(const Obj &SyclObject);
135  template <class T>
136  friend T sycl::detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
137 
138  std::shared_ptr<detail::node_impl> impl;
139 };
140 
141 namespace property {
142 namespace graph {
143 
146 class no_cycle_check : public ::sycl::detail::DataLessProperty<
147  ::sycl::detail::GraphNoCycleCheck> {
148 public:
149  no_cycle_check() = default;
150 };
151 
157  : public ::sycl::detail::DataLessProperty<
158  ::sycl::detail::GraphAssumeBufferOutlivesGraph> {
159 public:
161 };
162 
166  : public ::sycl::detail::DataLessProperty<::sycl::detail::GraphUpdatable> {
167 public:
168  updatable() = default;
169 };
170 
173 class enable_profiling : public ::sycl::detail::DataLessProperty<
174  ::sycl::detail::GraphEnableProfiling> {
175 public:
176  enable_profiling() = default;
177 };
178 } // namespace graph
179 
180 namespace node {
181 
184 class depends_on : public ::sycl::detail::PropertyWithData<
185  ::sycl::detail::GraphNodeDependencies> {
186 public:
187  template <typename... NodeTN> depends_on(NodeTN... nodes) : MDeps{nodes...} {}
188 
189  const std::vector<::sycl::ext::oneapi::experimental::node> &
191  return MDeps;
192  }
193 
194 private:
195  const std::vector<::sycl::ext::oneapi::experimental::node> MDeps;
196 };
197 
200 class depends_on_all_leaves : public ::sycl::detail::DataLessProperty<
201  ::sycl::detail::GraphDependOnAllLeaves> {
202 public:
204 };
205 
206 } // namespace node
207 } // namespace property
208 
209 namespace detail {
210 // Templateless modifiable command-graph base class.
211 class __SYCL_EXPORT modifiable_command_graph {
212 public:
217  modifiable_command_graph(const context &SyclContext, const device &SyclDevice,
218  const property_list &PropList = {});
219 
223  modifiable_command_graph(const queue &SyclQueue,
224  const property_list &PropList = {});
225 
229  node add(const property_list &PropList = {}) {
230  if (PropList.has_property<property::node::depends_on>()) {
231  auto Deps = PropList.get_property<property::node::depends_on>();
232  node Node = addImpl(Deps.get_dependencies());
233  if (PropList.has_property<property::node::depends_on_all_leaves>()) {
234  addGraphLeafDependencies(Node);
235  }
236  return Node;
237  }
238  node Node = addImpl({});
239  if (PropList.has_property<property::node::depends_on_all_leaves>()) {
240  addGraphLeafDependencies(Node);
241  }
242  return Node;
243  }
244 
249  template <typename T> node add(T CGF, const property_list &PropList = {}) {
250  if (PropList.has_property<property::node::depends_on>()) {
251  auto Deps = PropList.get_property<property::node::depends_on>();
252  node Node = addImpl(CGF, Deps.get_dependencies());
253  if (PropList.has_property<property::node::depends_on_all_leaves>()) {
254  addGraphLeafDependencies(Node);
255  }
256  return Node;
257  }
258  node Node = addImpl(CGF, {});
259  if (PropList.has_property<property::node::depends_on_all_leaves>()) {
260  addGraphLeafDependencies(Node);
261  }
262  return Node;
263  }
264 
268  void make_edge(node &Src, node &Dest);
269 
273  command_graph<graph_state::executable>
274  finalize(const property_list &PropList = {}) const;
275 
281  void begin_recording(queue &RecordingQueue,
282  const property_list &PropList = {});
283 
289  void begin_recording(const std::vector<queue> &RecordingQueues,
290  const property_list &PropList = {});
291 
293  void end_recording();
294 
297  void end_recording(queue &RecordingQueue);
298 
302  void end_recording(const std::vector<queue> &RecordingQueues);
303 
310  void print_graph(const std::string path, bool verbose = false) const;
311 
313  std::vector<node> get_nodes() const;
314 
316  std::vector<node> get_root_nodes() const;
317 
318 protected:
321  modifiable_command_graph(const std::shared_ptr<detail::graph_impl> &Impl)
322  : impl(Impl) {}
323 
328  node addImpl(std::function<void(handler &)> CGF,
329  const std::vector<node> &Dep);
330 
334  node addImpl(const std::vector<node> &Dep);
335 
339  void addGraphLeafDependencies(node Node);
340 
341  template <class Obj>
342  friend decltype(Obj::impl)
343  sycl::detail::getSyclObjImpl(const Obj &SyclObject);
344  template <class T>
345  friend T sycl::detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
346 
347  std::shared_ptr<detail::graph_impl> impl;
348 };
349 
350 // Templateless executable command-graph base class.
351 class __SYCL_EXPORT executable_command_graph {
352 public:
355 
358  void update(const command_graph<graph_state::modifiable> &Graph);
359 
363  void update(const node &Node);
364 
368  void update(const std::vector<node> &Nodes);
369 
370 protected:
375  executable_command_graph(const std::shared_ptr<detail::graph_impl> &Graph,
376  const sycl::context &Ctx,
377  const property_list &PropList = {});
378 
379  template <class Obj>
380  friend decltype(Obj::impl)
381  sycl::detail::getSyclObjImpl(const Obj &SyclObject);
382 
384  void finalizeImpl();
385 
386  std::shared_ptr<detail::exec_graph_impl> impl;
387 };
388 } // namespace detail
389 
391 template <graph_state State = graph_state::modifiable>
392 class command_graph : public detail::modifiable_command_graph {
393 public:
398  command_graph(const context &SyclContext, const device &SyclDevice,
399  const property_list &PropList = {})
400  : modifiable_command_graph(SyclContext, SyclDevice, PropList) {}
401 
405  command_graph(const queue &SyclQueue, const property_list &PropList = {})
406  : modifiable_command_graph(SyclQueue, PropList) {}
407 
408 private:
411  command_graph(const std::shared_ptr<detail::graph_impl> &Impl)
412  : modifiable_command_graph(Impl) {}
413 
414  template <class T>
415  friend T sycl::detail::createSyclObjFromImpl(decltype(T::impl) ImplObj);
416 };
417 
418 template <>
421 protected:
425 };
426 
427 namespace detail {
428 class __SYCL_EXPORT dynamic_parameter_base {
429 public:
432  Graph,
433  size_t ParamSize, const void *Data);
434 
435 protected:
436  void updateValue(const void *NewValue, size_t Size);
437 
438  void updateAccessor(const sycl::detail::AccessorBaseHost *Acc);
439  std::shared_ptr<dynamic_parameter_impl> impl;
440 
441  template <class Obj>
442  friend decltype(Obj::impl)
443  sycl::detail::getSyclObjImpl(const Obj &SyclObject);
444 };
445 } // namespace detail
446 
447 template <typename ValueT>
448 class dynamic_parameter : public detail::dynamic_parameter_base {
449  static constexpr bool IsAccessor =
450  std::is_base_of_v<sycl::detail::AccessorBaseHost, ValueT>;
451  static constexpr sycl::detail::kernel_param_kind_t ParamType =
452  IsAccessor ? sycl::detail::kernel_param_kind_t::kind_accessor
453  : std::is_pointer_v<ValueT>
454  ? sycl::detail::kernel_param_kind_t::kind_pointer
455  : sycl::detail::kernel_param_kind_t::kind_std_layout;
456 
457 public:
462  const ValueT &Param)
463  : detail::dynamic_parameter_base(Graph, sizeof(ValueT), &Param) {}
464 
467  void update(const ValueT &NewValue) {
468  if constexpr (IsAccessor) {
470  } else {
471  detail::dynamic_parameter_base::updateValue(&NewValue, sizeof(ValueT));
472  }
473  }
474 };
475 
477 template <typename ValueT>
479  const ValueT &Param) -> dynamic_parameter<ValueT>;
480 template <graph_state State = graph_state::modifiable>
481 command_graph(const context &SyclContext, const device &SyclDevice,
482  const property_list &PropList) -> command_graph<State>;
483 
484 } // namespace experimental
485 } // namespace oneapi
486 } // namespace ext
487 
488 template <>
489 struct is_property<ext::oneapi::experimental::property::graph::no_cycle_check>
490  : std::true_type {};
491 
492 template <>
493 struct is_property<ext::oneapi::experimental::property::node::depends_on>
494  : std::true_type {};
495 
496 template <>
498  ext::oneapi::experimental::property::graph::no_cycle_check,
500  ext::oneapi::experimental::graph_state::modifiable>> : std::true_type {
501 };
502 
503 template <>
504 struct is_property_of<ext::oneapi::experimental::property::node::depends_on,
505  ext::oneapi::experimental::node> : std::true_type {};
506 
507 } // namespace _V1
508 } // namespace sycl
The file contains implementations of accessor class.
The context class represents a SYCL context on which kernel functions may be executed.
Definition: context.hpp:50
The SYCL device class encapsulates a single SYCL device on which kernels may be executed.
Definition: device.hpp:64
An event object can be used to synchronize memory transfers, enqueues of kernels and signaling barrie...
Definition: event.hpp:44
command_graph(const queue &SyclQueue, const property_list &PropList={})
Constructor.
Definition: graph.hpp:405
command_graph(const context &SyclContext, const device &SyclDevice, const property_list &PropList={})
Constructor.
Definition: graph.hpp:398
void updateAccessor(const sycl::detail::AccessorBaseHost *Acc)
std::shared_ptr< dynamic_parameter_impl > impl
Definition: graph.hpp:439
Class representing the implementation of command_graph<executable>.
executable_command_graph()=delete
An executable command-graph is not user constructable.
Implementation details of command_graph<modifiable>.
Definition: graph_impl.hpp:849
command_graph< graph_state::executable > finalize(const property_list &PropList={}) const
Finalize modifiable graph into an executable graph.
void begin_recording(queue &RecordingQueue, const property_list &PropList={})
Change the state of a queue to be recording and associate this graph with it.
void begin_recording(const std::vector< queue > &RecordingQueues, const property_list &PropList={})
Change the state of multiple queues to be recording and associate this graph with each of them.
modifiable_command_graph(const std::shared_ptr< detail::graph_impl > &Impl)
Constructor used internally by the runtime.
Definition: graph.hpp:321
node add(const property_list &PropList={})
Add an empty node to the graph.
Definition: graph.hpp:229
modifiable_command_graph(const context &SyclContext, const device &SyclDevice, const property_list &PropList={})
Constructor.
modifiable_command_graph(const queue &SyclQueue, const property_list &PropList={})
Constructor.
node add(T CGF, const property_list &PropList={})
Add a command-group node to the graph.
Definition: graph.hpp:249
void update(const ValueT &NewValue)
Updates this dynamic parameter and all registered nodes with a new value.
Definition: graph.hpp:467
dynamic_parameter(experimental::command_graph< graph_state::modifiable > Graph, const ValueT &Param)
Constructs a new dynamic parameter.
Definition: graph.hpp:461
Class representing a node in the graph, returned by command_graph::add().
Definition: graph.hpp:105
void update_nd_range(nd_range< Dimensions > executionRange)
Update the ND-Range of this node if it is a kernel execution node.
void update_range(range< Dimensions > executionRange)
Update the Range of this node if it is a kernel execution node.
Property passed to command_graph constructor to allow buffers to be used with graphs.
Definition: graph.hpp:158
Property used to enable executable graph profiling.
Definition: graph.hpp:174
Property passed to command_graph constructor to disable checking for cycles.
Definition: graph.hpp:147
Property passed to command_graph<graph_state::modifiable>::finalize() to mark the resulting executabl...
Definition: graph.hpp:166
Property used to to add all previous graph leaves as dependencies when creating a new node with comma...
Definition: graph.hpp:201
Property used to define dependent nodes when creating a new node with command_graph::add().
Definition: graph.hpp:185
const std::vector<::sycl::ext::oneapi::experimental::node > & get_dependencies() const
Definition: graph.hpp:190
Command group handler class.
Definition: handler.hpp:462
Provides an abstraction of a SYCL kernel.
Definition: kernel.hpp:76
Defines the iteration domain of both the work-groups and the overall dispatch.
Definition: nd_range.hpp:22
Objects of the property_list class are containers for the SYCL properties.
Encapsulates a single SYCL queue which schedules kernels on a SYCL device.
Definition: queue.hpp:111
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: impl_utils.hpp:30
T createSyclObjFromImpl(decltype(T::impl) ImplObj)
Definition: impl_utils.hpp:48
const char * UnsupportedFeatureToString(UnsupportedGraphFeatures Feature)
Definition: graph.hpp:60
graph_state
State to template the command_graph class on.
Definition: graph.hpp:37
@ modifiable
In modifiable state, commands can be added to graph.
@ executable
In executable state, the graph is ready to execute.
command_graph(const context &SyclContext, const device &SyclDevice, const property_list &PropList) -> command_graph< State >
dynamic_parameter(experimental::command_graph< graph_state::modifiable > Graph, const ValueT &Param) -> dynamic_parameter< ValueT >
Additional CTAD deduction guides.
Definition: access.hpp:18