Skip to content

[SYCL][Doc] Update spec example with buffer restrictions #304

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
EwanC merged 1 commit into from
Aug 22, 2023
Merged

[SYCL][Doc] Update spec example with buffer restrictions #304

Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
44 changes: 30 additions & 14 deletions sycl/doc/extensions/experimental/sycl_ext_oneapi_graph.asciidoc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -476,7 +476,7 @@ node.

[source,c++]
----
namespace sycl::ext::oneapi::experimental::property::node
namespace sycl::ext::oneapi::experimental::property::node {
class depends_on {
public:
template<typename... NodeTN>
Expand Down Expand Up @@ -1376,14 +1376,24 @@ submitted in its entirety for execution via
[source, c++]
----
using namespace sycl;
namespace sycl_ext = sycl::ext::oneapi::experimental;

queue q{default_selector{}};

// New object representing graph of command-groups
ext::oneapi::experimental::command_graph graph(q.get_context(), q.get_device());
// Lifetime of buffers must exceed the lifetime of graphs they are used in.
buffer<T> bufferA{dataA.data(), range<1>{elements}};
bufferA.set_write_back(false);
buffer<T> bufferB{dataB.data(), range<1>{elements}};
bufferB.set_write_back(false);
buffer<T> bufferC{dataC.data(), range<1>{elements}};
bufferC.set_write_back(false);

{
buffer<T> bufferA{dataA.data(), range<1>{elements}};
buffer<T> bufferB{dataB.data(), range<1>{elements}};
buffer<T> bufferC{dataC.data(), range<1>{elements}};
// New object representing graph of command-groups
sycl_ext::command_graph graph(q.get_context(), q.get_device(),
{sycl_ext::property::graph::assume_buffer_outlives_graph{},
sycl_ext::property::graph::assume_data_outlives_buffer{}});


// `q` will be put in the recording state where commands are recorded to
// `graph` rather than submitted for execution immediately.
Expand Down Expand Up @@ -1433,17 +1443,23 @@ submitted in its entirety for execution via
// queue `q` will be returned to the executing state where commands are
// submitted immediately for extension.
graph.end_recording();
}

// Finalize the modifiable graph to create an executable graph that can be
// submitted for execution.
auto exec_graph = graph.finalize();
// Finalize the modifiable graph to create an executable graph that can be
// submitted for execution.
auto exec_graph = graph.finalize();

// Execute graph
q.submit([&](handler& cgh) {
cgh.ext_oneapi_graph(exec_graph);
});
// Execute graph
q.submit([&](handler& cgh) {
cgh.ext_oneapi_graph(exec_graph);
}).wait();
}

// Check output using host accessors
host_accessor hostAccA(bufferA);
host_accessor hostAccB(bufferB);
host_accessor hostAccC(bufferC);

...
----

== Future Direction [[future-direction]]
Expand Down