forked from intel/llvm
-
Notifications
You must be signed in to change notification settings - Fork 4
Commit
This commit does not belong to any branch on this repository, and may belong to a fork outside of the repository.
[SYCL][Graph] command_graph queue constructor
Feedback from Jack Kirk (@JackAKirk) that SYCL classes with a sycl::context and sycl::device constructor normally have a matching constructor taking a sycl::queue which replaces the context and device. Being able to use a queue constructor makes it easier to port code from CUDA/HIP where there is not a concept analogous to SYCL contexts. Our tests/examples also commonly use the pattern ```cpp queue Queue; command_graph Graph {Queue.get_context(), Queue.get_device()} ``` So being able to use a queue constructor is also a more concise way to call the constructor in a lot of cases, regardless of platform.
- Loading branch information
Showing
11 changed files
with
225 additions
and
2 deletions.
There are no files selected for viewing
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,13 @@ | ||
// RUN: %{build} -o %t.out | ||
// RUN: %{run} %t.out | ||
// Extra run to check for leaks in Level Zero using UR_L0_LEAKS_DEBUG | ||
// RUN: %if level_zero %{env UR_L0_LEAKS_DEBUG=1 %{run} %t.out 2>&1 | FileCheck %s %} | ||
// | ||
// CHECK-NOT: LEAK | ||
// | ||
// TODO enable cuda once buffer issue investigated and fixed | ||
// UNSUPPORTED: cuda | ||
|
||
#define GRAPH_E2E_EXPLICIT | ||
|
||
#include "../Inputs/queue_constructor_buffer.cpp" |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,10 @@ | ||
// RUN: %{build} -o %t.out | ||
// RUN: %{run} %t.out | ||
// Extra run to check for leaks in Level Zero using UR_L0_LEAKS_DEBUG | ||
// RUN: %if level_zero %{env UR_L0_LEAKS_DEBUG=1 %{run} %t.out 2>&1 | FileCheck %s %} | ||
// | ||
// CHECK-NOT: LEAK | ||
|
||
#define GRAPH_E2E_EXPLICIT | ||
|
||
#include "../Inputs/queue_constructor_usm.cpp" |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,59 @@ | ||
// Tests a graph created with the simplified sycl::queue constructor works | ||
// as expected. | ||
|
||
#include "../graph_common.hpp" | ||
|
||
int main() { | ||
queue Queue{{sycl::ext::intel::property::queue::no_immediate_command_list{}}}; | ||
|
||
if (!are_graphs_supported(Queue)) { | ||
return 0; | ||
} | ||
|
||
using T = unsigned short; | ||
|
||
std::vector<T> DataA(Size), DataB(Size), DataC(Size); | ||
|
||
std::iota(DataA.begin(), DataA.end(), 1); | ||
std::iota(DataB.begin(), DataB.end(), 10); | ||
std::iota(DataC.begin(), DataC.end(), 1000); | ||
|
||
std::vector<T> ReferenceA(DataA), ReferenceB(DataB), ReferenceC(DataC); | ||
calculate_reference_data(Iterations, Size, ReferenceA, ReferenceB, | ||
ReferenceC); | ||
|
||
buffer<T> BufferA{DataA.data(), range<1>{DataA.size()}}; | ||
BufferA.set_write_back(false); | ||
buffer<T> BufferB{DataB.data(), range<1>{DataB.size()}}; | ||
BufferB.set_write_back(false); | ||
buffer<T> BufferC{DataC.data(), range<1>{DataC.size()}}; | ||
BufferC.set_write_back(false); | ||
{ | ||
exp_ext::command_graph Graph{ | ||
Queue, {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; | ||
|
||
// Add commands to graph | ||
add_nodes(Graph, Queue, Size, BufferA, BufferB, BufferC); | ||
|
||
auto GraphExec = Graph.finalize(); | ||
|
||
event Event; | ||
for (unsigned n = 0; n < Iterations; n++) { | ||
Event = | ||
Queue.submit([&](handler &CGH) { CGH.ext_oneapi_graph(GraphExec); }); | ||
} | ||
Queue.wait_and_throw(); | ||
} | ||
|
||
host_accessor HostAccA(BufferA); | ||
host_accessor HostAccB(BufferB); | ||
host_accessor HostAccC(BufferC); | ||
|
||
for (size_t i = 0; i < Size; i++) { | ||
assert(check_value(i, ReferenceA[i], HostAccA[i], "HostAccA")); | ||
assert(check_value(i, ReferenceB[i], HostAccB[i], "HostAccB")); | ||
assert(check_value(i, ReferenceC[i], HostAccC[i], "HostAccC")); | ||
} | ||
|
||
return 0; | ||
} |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,68 @@ | ||
// Tests a graph created with the simplified sycl::queue constructor works | ||
// as expected. | ||
|
||
#include "../graph_common.hpp" | ||
|
||
int main() { | ||
queue Queue{{sycl::ext::intel::property::queue::no_immediate_command_list{}}}; | ||
|
||
if (!are_graphs_supported(Queue)) { | ||
return 0; | ||
} | ||
|
||
using T = int; | ||
|
||
std::vector<T> DataA(Size), DataB(Size), DataC(Size); | ||
std::iota(DataA.begin(), DataA.end(), 1); | ||
std::iota(DataB.begin(), DataB.end(), 10); | ||
std::iota(DataC.begin(), DataC.end(), 1000); | ||
|
||
std::vector<T> ReferenceA(DataA), ReferenceB(DataB), ReferenceC(DataC); | ||
calculate_reference_data(Iterations, Size, ReferenceA, ReferenceB, | ||
ReferenceC); | ||
|
||
exp_ext::command_graph Graph{Queue}; | ||
|
||
T *PtrA = malloc_device<T>(Size, Queue); | ||
T *PtrB = malloc_device<T>(Size, Queue); | ||
T *PtrC = malloc_device<T>(Size, Queue); | ||
|
||
Queue.copy(DataA.data(), PtrA, Size); | ||
Queue.copy(DataB.data(), PtrB, Size); | ||
Queue.copy(DataC.data(), PtrC, Size); | ||
Queue.wait_and_throw(); | ||
|
||
// Add commands to graph | ||
add_nodes(Graph, Queue, Size, PtrA, PtrB, PtrC); | ||
|
||
auto GraphExec = Graph.finalize(); | ||
|
||
auto SubmitGraph = [&]() { | ||
Queue.submit([&](handler &CGH) { CGH.ext_oneapi_graph(GraphExec); }); | ||
}; | ||
|
||
event Event; | ||
for (unsigned n = 0; n < Iterations; n++) { | ||
Event = | ||
Queue.submit([&](handler &CGH) { CGH.ext_oneapi_graph(GraphExec); }); | ||
} | ||
|
||
Queue.wait_and_throw(); | ||
|
||
Queue.copy(PtrA, DataA.data(), Size); | ||
Queue.copy(PtrB, DataB.data(), Size); | ||
Queue.copy(PtrC, DataC.data(), Size); | ||
Queue.wait_and_throw(); | ||
|
||
free(PtrA, Queue); | ||
free(PtrB, Queue); | ||
free(PtrC, Queue); | ||
|
||
for (size_t i = 0; i < Size; i++) { | ||
assert(check_value(i, ReferenceA[i], DataA[i], "DataA")); | ||
assert(check_value(i, ReferenceB[i], DataB[i], "DataB")); | ||
assert(check_value(i, ReferenceC[i], DataC[i], "DataC")); | ||
} | ||
|
||
return 0; | ||
} |
13 changes: 13 additions & 0 deletions
13
sycl/test-e2e/Graph/RecordReplay/queue_constructor_buffer.cpp
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,13 @@ | ||
// RUN: %{build} -o %t.out | ||
// RUN: %{run} %t.out | ||
// Extra run to check for leaks in Level Zero using UR_L0_LEAKS_DEBUG | ||
// RUN: %if level_zero %{env UR_L0_LEAKS_DEBUG=1 %{run} %t.out 2>&1 | FileCheck %s %} | ||
// | ||
// CHECK-NOT: LEAK | ||
// | ||
// TODO enable cuda once buffer issue investigated and fixed | ||
// UNSUPPORTED: cuda | ||
|
||
#define GRAPH_E2E_RECORD_REPLAY | ||
|
||
#include "../Inputs/queue_constructor_buffer.cpp" |
10 changes: 10 additions & 0 deletions
10
sycl/test-e2e/Graph/RecordReplay/queue_constructor_usm.cpp
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Original file line number | Diff line number | Diff line change |
---|---|---|
@@ -0,0 +1,10 @@ | ||
// RUN: %{build} -o %t.out | ||
// RUN: %{run} %t.out | ||
// Extra run to check for leaks in Level Zero using UR_L0_LEAKS_DEBUG | ||
// RUN: %if level_zero %{env UR_L0_LEAKS_DEBUG=1 %{run} %t.out 2>&1 | FileCheck %s %} | ||
// | ||
// CHECK-NOT: LEAK | ||
|
||
#define GRAPH_E2E_RECORD_REPLAY | ||
|
||
#include "../Inputs/queue_constructor_usm.cpp" |
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters