DPC++ Runtime
Runtime libraries for oneAPI DPC++
stream_impl.cpp
Go to the documentation of this file.
1 //==----------------- stream_impl.cpp - SYCL standard header file ---------==//
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 #include <detail/queue_impl.hpp>
11 #include <detail/stream_impl.hpp>
12 #include <sycl/queue.hpp>
13 
14 #include <cstdio>
15 
16 namespace sycl {
18 namespace detail {
19 
20 stream_impl::stream_impl(size_t BufferSize, size_t MaxStatementSize,
21  handler &CGH)
22  : stream_impl(BufferSize, MaxStatementSize, {}) {
23  (void)CGH;
24 }
25 
26 stream_impl::stream_impl(size_t BufferSize, size_t MaxStatementSize,
27  const property_list &PropList)
28  : BufferSize_(BufferSize), MaxStatementSize_(MaxStatementSize),
29  PropList_(PropList) {
30  // We need to store stream buffers in the scheduler because they need to be
31  // alive after submitting the kernel. They cannot be stored in the stream
32  // object because it causes loop dependency between objects and results in
33  // memory leak.
34  // Allocate additional place in the stream buffer for the offset variable and
35  // the end of line symbol.
37  this, BufferSize + OffsetSize + 1 /* size of the stream buffer */,
38  MaxStatementSize + FLUSH_BUF_OFFSET_SIZE /* size of the flush buffer */);
39 }
40 
41 // Method to provide an access to the global stream buffer
44  .StreamBuffersPool.find(this)
45  ->second->Buf.get_access<sycl::access::mode::read_write>(
46  CGH, range<1>(BufferSize_), id<1>(OffsetSize));
47 }
48 
49 // Method to provide an accessor to the global flush buffer
52  .StreamBuffersPool.find(this)
53  ->second->FlushBuf.get_access<sycl::access::mode::read_write>(
54  CGH, range<1>(MaxStatementSize_ + FLUSH_BUF_OFFSET_SIZE), id<1>(0));
55 }
56 
57 // Method to provide an atomic access to the offset in the global stream
58 // buffer and offset in the flush buffer
61  .StreamBuffersPool.find(this)
62  ->second->Buf,
63  id<1>(0), range<1>(OffsetSize));
64  auto ReinterpretedBuf = OffsetSubBuf.reinterpret<unsigned, 1>(range<1>(2));
65  return ReinterpretedBuf.get_access<sycl::access::mode::atomic>(
66  CGH, range<1>(2), id<1>(0));
67 }
68 size_t stream_impl::get_size() const { return BufferSize_; }
69 
70 size_t stream_impl::get_max_statement_size() const { return MaxStatementSize_; }
71 
72 void stream_impl::flush(const EventImplPtr &LeadEvent) {
73  // We don't want stream flushing to be blocking operation that is why submit a
74  // host task to print stream buffer. It will fire up as soon as the kernel
75  // finishes execution.
76  auto Q = detail::createSyclObjFromImpl<queue>(
77  sycl::detail::Scheduler::getInstance().getDefaultHostQueue());
78  event Event = Q.submit([&](handler &cgh) {
79  auto BufHostAcc =
81  .StreamBuffersPool.find(this)
82  ->second->Buf
84  cgh, range<1>(BufferSize_), id<1>(OffsetSize));
85  // Create accessor to the flush buffer even if not using it yet. Otherwise
86  // kernel will be a leaf for the flush buffer and scheduler will not be able
87  // to cleanup the kernel. TODO: get rid of finalize method by using host
88  // accessor to the flush buffer.
89  auto FlushBufHostAcc =
91  .StreamBuffersPool.find(this)
92  ->second->FlushBuf
94  cgh);
95  cgh.host_task([=] {
96  printf("%s", &(BufHostAcc[0]));
97  fflush(stdout);
98  });
99  });
100  if (LeadEvent) {
101  LeadEvent->attachEventToComplete(detail::getSyclObjImpl(Event));
102  LeadEvent->getSubmittedQueue()->registerStreamServiceEvent(
103  detail::getSyclObjImpl(Event));
104  }
105 }
106 
107 void stream_impl::flush() { flush(nullptr); }
108 } // namespace detail
109 } // __SYCL_INLINE_VER_NAMESPACE(_V1)
110 } // namespace sycl
void allocateStreamBuffers(stream_impl *, size_t, size_t)
Allocate buffers in the pool for a provided stream.
Definition: scheduler.cpp:369
static Scheduler & getInstance()
Definition: scheduler.cpp:207
std::unordered_map< stream_impl *, StreamBuffers * > StreamBuffersPool
Definition: scheduler.hpp:846
size_t get_max_statement_size() const
Definition: stream_impl.cpp:70
GlobalBufAccessorT accessGlobalFlushBuf(handler &CGH)
Definition: stream_impl.cpp:50
stream_impl(size_t BufferSize, size_t MaxStatementSize, handler &CGH)
Definition: stream_impl.cpp:20
GlobalOffsetAccessorT accessGlobalOffset(handler &CGH)
Definition: stream_impl.cpp:59
GlobalBufAccessorT accessGlobalBuf(handler &CGH)
Definition: stream_impl.cpp:42
Command group handler class.
Definition: handler.hpp:310
Objects of the property_list class are containers for the SYCL properties.
#define __SYCL_INLINE_VER_NAMESPACE(X)
decltype(Obj::impl) getSyclObjImpl(const Obj &SyclObject)
Definition: common.hpp:240
std::shared_ptr< event_impl > EventImplPtr
Definition: cg.hpp:42
constexpr unsigned FLUSH_BUF_OFFSET_SIZE
Definition: stream.hpp:58
int printf(const FormatT *__format, Args... args)
Definition: builtins.hpp:86
---— Error handling, matching OpenCL plugin semantics.
Definition: access.hpp:14