From 5b8952777ce0bebc4943f69bcfbc56e703dde421 Mon Sep 17 00:00:00 2001 From: Ben Tracy Date: Wed, 6 Mar 2024 18:14:16 +0000 Subject: [PATCH] Fix inconsistencies between spec and implementation --- .../sycl/ext/oneapi/experimental/graph.hpp | 17 +++++++++++------ .../Graph/Explicit/update_before_finalize.cpp | 2 +- .../Explicit/update_with_indices_accessor.cpp | 2 +- ...update_with_indices_multiple_exec_graphs.cpp | 2 +- .../Explicit/update_with_indices_ordering.cpp | 2 +- .../Graph/Explicit/update_with_indices_ptr.cpp | 2 +- .../update_with_indices_ptr_double_update.cpp | 2 +- .../update_with_indices_ptr_multiple_nodes.cpp | 2 +- .../update_with_indices_ptr_multiple_params.cpp | 6 +++--- .../update_with_indices_ptr_subgraph.cpp | 2 +- .../Explicit/update_with_indices_scalar.cpp | 2 +- .../Extensions/CommandGraph/Update.cpp | 4 ++-- 12 files changed, 25 insertions(+), 20 deletions(-) diff --git a/sycl/include/sycl/ext/oneapi/experimental/graph.hpp b/sycl/include/sycl/ext/oneapi/experimental/graph.hpp index 057bb12db2be0..a4610f8591f02 100644 --- a/sycl/include/sycl/ext/oneapi/experimental/graph.hpp +++ b/sycl/include/sycl/ext/oneapi/experimental/graph.hpp @@ -452,13 +452,15 @@ class dynamic_parameter : public detail::dynamic_parameter_base { : sycl::detail::kernel_param_kind_t::kind_std_layout; public: - dynamic_parameter(experimental::command_graph Graph) + /// Constructs a new dynamic parameter. + /// @param Graph The graph associated with this parameter. + /// @param Param A reference value for this parameter used for CTAD. + dynamic_parameter(experimental::command_graph Graph, + const ValueT &Param) : detail::dynamic_parameter_base(Graph), MValue() {} - dynamic_parameter(ValueT InitialValue, - experimental::command_graph Graph) - : detail::dynamic_parameter_base(Graph), MValue(InitialValue) {} - + /// Updates this dynamic parameter and all registered nodes with a new value. + /// @param NewValue The new value for the parameter. void update(const ValueT &NewValue) { MValue = NewValue; if constexpr (IsAccessor) { @@ -472,7 +474,10 @@ class dynamic_parameter : public detail::dynamic_parameter_base { ValueT MValue; }; -/// Additional CTAD deduction guide. +/// Additional CTAD deduction guides. +template +dynamic_parameter(experimental::command_graph Graph, + const ValueT &Param) -> dynamic_parameter; template command_graph(const context &SyclContext, const device &SyclDevice, const property_list &PropList) -> command_graph; diff --git a/sycl/test-e2e/Graph/Explicit/update_before_finalize.cpp b/sycl/test-e2e/Graph/Explicit/update_before_finalize.cpp index a1827e56de3fa..c2d5af7735c14 100644 --- a/sycl/test-e2e/Graph/Explicit/update_before_finalize.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_before_finalize.cpp @@ -31,7 +31,7 @@ int main() { Queue.memset(PtrA, 0, N * sizeof(int)).wait(); Queue.memset(PtrB, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(Graph); + exp_ext::dynamic_parameter InputParam(Graph, PtrA); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_accessor.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_accessor.cpp index ed317c97a5085..6ea48246ee1fa 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_accessor.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_accessor.cpp @@ -32,7 +32,7 @@ int main() { BufA.set_write_back(false); BufB.set_write_back(false); - exp_ext::dynamic_parameter InputParam(BufA.get_access(), Graph); + exp_ext::dynamic_parameter InputParam(Graph, BufA.get_access()); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_multiple_exec_graphs.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_multiple_exec_graphs.cpp index 0764327407da1..c3f730d8d1831 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_multiple_exec_graphs.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_multiple_exec_graphs.cpp @@ -32,7 +32,7 @@ int main() { Queue.memset(PtrA, 0, N * sizeof(int)).wait(); Queue.memset(PtrB, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(Graph); + exp_ext::dynamic_parameter InputParam(Graph, PtrA); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_ordering.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_ordering.cpp index 9f976f8bbb56f..63da9510d264d 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_ordering.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_ordering.cpp @@ -36,7 +36,7 @@ int main() { Queue.memset(PtrA, 0, N * sizeof(int)).wait(); Queue.memset(PtrB, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(Graph); + exp_ext::dynamic_parameter InputParam(Graph, PtrA); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr.cpp index 838c578358999..127543cdcb400 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr.cpp @@ -31,7 +31,7 @@ int main() { Queue.memset(PtrA, 0, N * sizeof(int)).wait(); Queue.memset(PtrB, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(Graph); + exp_ext::dynamic_parameter InputParam(Graph, PtrA); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_double_update.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_double_update.cpp index b1443f62a33b3..fd7ec939f90c0 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_double_update.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_double_update.cpp @@ -34,7 +34,7 @@ int main() { Queue.memset(PtrB, 0, N * sizeof(int)).wait(); Queue.memset(PtrUnused, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(Graph); + exp_ext::dynamic_parameter InputParam(Graph, PtrA); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_nodes.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_nodes.cpp index b725bd137efa1..42b92d4ecd02b 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_nodes.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_nodes.cpp @@ -32,7 +32,7 @@ int main() { Queue.memset(PtrA, 0, N * sizeof(int)).wait(); Queue.memset(PtrB, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(Graph); + exp_ext::dynamic_parameter InputParam(Graph, PtrA); auto KernelNodeA = Graph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_params.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_params.cpp index 7e62d8be9b3f8..4eb411e1cb6e5 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_params.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_multiple_params.cpp @@ -39,9 +39,9 @@ int main() { Queue.memcpy(PtrB, HostDataB.data(), N * sizeof(int)).wait(); Queue.memset(PtrC, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter ParamA(PtrA, Graph); - exp_ext::dynamic_parameter ParamB(PtrB, Graph); - exp_ext::dynamic_parameter ParamOut(PtrC, Graph); + exp_ext::dynamic_parameter ParamA(Graph, PtrA); + exp_ext::dynamic_parameter ParamB(Graph, PtrB); + exp_ext::dynamic_parameter ParamOut(Graph, PtrC); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input pointers, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_subgraph.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_subgraph.cpp index 6c068445193c4..386d503e4b0b5 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_subgraph.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_ptr_subgraph.cpp @@ -34,7 +34,7 @@ int main() { Queue.memset(PtrA, 0, N * sizeof(int)).wait(); Queue.memset(PtrB, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(SubGraph); + exp_ext::dynamic_parameter InputParam(SubGraph, PtrA); auto SubKernelNode = SubGraph.add([&](handler &cgh) { // Register the input pointer, we should be using set_arg but can't diff --git a/sycl/test-e2e/Graph/Explicit/update_with_indices_scalar.cpp b/sycl/test-e2e/Graph/Explicit/update_with_indices_scalar.cpp index 3ffa2e3b0b9a2..0ee594b005293 100644 --- a/sycl/test-e2e/Graph/Explicit/update_with_indices_scalar.cpp +++ b/sycl/test-e2e/Graph/Explicit/update_with_indices_scalar.cpp @@ -30,7 +30,7 @@ int main() { Queue.memset(DeviceData, 0, N * sizeof(int)).wait(); - exp_ext::dynamic_parameter InputParam(ScalarValue, Graph); + exp_ext::dynamic_parameter InputParam(Graph, ScalarValue); auto KernelNode = Graph.add([&](handler &cgh) { // Register the input scalar, we should be using set_arg but can't diff --git a/sycl/unittests/Extensions/CommandGraph/Update.cpp b/sycl/unittests/Extensions/CommandGraph/Update.cpp index 7d1e1b8cfa8f7..72ea40f2a8e97 100644 --- a/sycl/unittests/Extensions/CommandGraph/Update.cpp +++ b/sycl/unittests/Extensions/CommandGraph/Update.cpp @@ -29,7 +29,7 @@ TEST_F(CommandGraphTest, UpdatableException) { TEST_F(CommandGraphTest, DynamicParamRegister) { // Check that registering a dynamic param with a node from a graph that was // not passed to its constructor throws. - experimental::dynamic_parameter DynamicParam(Graph); + experimental::dynamic_parameter DynamicParam(Graph, int{}); auto OtherGraph = experimental::command_graph(Queue.get_context(), Queue.get_device()); @@ -72,7 +72,7 @@ TEST_F(CommandGraphTest, UpdateNodeTypeExceptions) { int *PtrA = malloc_device(16, Queue); int *PtrB = malloc_device(16, Queue); - experimental::dynamic_parameter DynamicParam{Graph}; + experimental::dynamic_parameter DynamicParam{Graph, int{}}; ASSERT_NO_THROW(auto NodeKernel = Graph.add([&](sycl::handler &cgh) { DynamicParam.register_with_node(cgh, 0);