From 45d9b7da6ce125072c65f88e785729f73545616d Mon Sep 17 00:00:00 2001 From: Matthew Michel Date: Fri, 10 Oct 2025 13:56:02 -0700 Subject: [PATCH 1/3] Add test for linear graph optimization --- sycl/source/detail/graph/graph_impl.hpp | 7 ++ .../Extensions/CommandGraph/CMakeLists.txt | 1 + .../Extensions/CommandGraph/Common.hpp | 23 +++- .../CommandGraph/LinearGraphOptimization.cpp | 109 ++++++++++++++++++ 4 files changed, 139 insertions(+), 1 deletion(-) create mode 100644 sycl/unittests/Extensions/CommandGraph/LinearGraphOptimization.cpp diff --git a/sycl/source/detail/graph/graph_impl.hpp b/sycl/source/detail/graph/graph_impl.hpp index e407accd50f41..04f4bef37eadf 100644 --- a/sycl/source/detail/graph/graph_impl.hpp +++ b/sycl/source/detail/graph/graph_impl.hpp @@ -23,6 +23,9 @@ #include // for shared_mutex #include // for vector +// For testing of graph internals +class GraphImplTest; + namespace sycl { inline namespace _V1 { // Forward declarations @@ -732,6 +735,10 @@ class exec_graph_impl { } private: + // Test helper class for inspecting private graph internals to validate + // under-the-hood behavior and optimizations. + friend class ::GraphImplTest; + /// Create a command-group for the node and add it to command-buffer by going /// through the scheduler. /// @param CommandBuffer Command-buffer to add node to as a command. diff --git a/sycl/unittests/Extensions/CommandGraph/CMakeLists.txt b/sycl/unittests/Extensions/CommandGraph/CMakeLists.txt index 928d67eb56112..70d50bdc4f5b1 100644 --- a/sycl/unittests/Extensions/CommandGraph/CMakeLists.txt +++ b/sycl/unittests/Extensions/CommandGraph/CMakeLists.txt @@ -13,4 +13,5 @@ add_sycl_unittest(CommandGraphExtensionTests OBJECT TopologicalSort.cpp Update.cpp Properties.cpp + LinearGraphOptimization.cpp ) diff --git a/sycl/unittests/Extensions/CommandGraph/Common.hpp b/sycl/unittests/Extensions/CommandGraph/Common.hpp index 5d0842f842a29..932f66c5b2283 100644 --- a/sycl/unittests/Extensions/CommandGraph/Common.hpp +++ b/sycl/unittests/Extensions/CommandGraph/Common.hpp @@ -24,9 +24,30 @@ using namespace sycl; using namespace sycl::ext::oneapi; - using sycl::detail::getSyclObjImpl; +// Implement the test friend class forward declared in graph_impl.hpp so tests +// can access private members to analyze internal optimizations (partitions, +// sync points). +class GraphImplTest { + using exec_graph_impl = + sycl::ext::oneapi::experimental::detail::exec_graph_impl; + using partition = sycl::ext::oneapi::experimental::detail::partition; + +public: + static int NumPartitionsInOrder(const exec_graph_impl &Impl) { + int NumInOrder = 0; + for (const auto &P : Impl.MPartitions) { + if (P && P->MIsInOrderGraph) + ++NumInOrder; + } + return NumInOrder; + } + static int NumSyncPoints(const exec_graph_impl &Impl) { + return Impl.MSyncPoints.size(); + } +}; + // Common Test fixture class CommandGraphTest : public ::testing::Test { public: diff --git a/sycl/unittests/Extensions/CommandGraph/LinearGraphOptimization.cpp b/sycl/unittests/Extensions/CommandGraph/LinearGraphOptimization.cpp new file mode 100644 index 0000000000000..20b9aa754464f --- /dev/null +++ b/sycl/unittests/Extensions/CommandGraph/LinearGraphOptimization.cpp @@ -0,0 +1,109 @@ +//==--------------------- LinearGraphOptimization.cpp ----------------------==// +// +// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. +// See https://llvm.org/LICENSE.txt for license information. +// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception +// +//===----------------------------------------------------------------------===// + +// Test for linear graph optimization which skips creating and tracking UR sync +// points. Optimization is an internal implementation detail, validated through +// inspecting private members of exec_graph_impl. Test achieves two goals: 1) +// Validates that linear partitions in graphs are optimized to avoid using UR +// sync points 2) Validates that non-linear partitions contain the expected +// number of sync points + +#include "Common.hpp" + +using namespace sycl; +using namespace sycl::ext::oneapi::experimental; +using namespace sycl::ext::oneapi::experimental::detail; + +// Helper to build a linear chain of N kernels on a queue inside graph capture. +static void BuildLinearChain(queue &Queue, bool IsInOrderQueue, int N) { + sycl::event Event{}; + for (int I = 0; I < N; ++I) { + Event = Queue.submit([&](handler &h) { + if (I > 0 && !IsInOrderQueue) { + h.depends_on(Event); + } + h.single_task([] {}); + }); + } +} + +// Validate linear optimization invariants on an executable graph. +static void ValidateLinearExec(exec_graph_impl &Impl, int NumLinearChains) { + EXPECT_EQ(GraphImplTest::NumPartitionsInOrder(Impl), NumLinearChains); + EXPECT_EQ(GraphImplTest::NumSyncPoints(Impl), 0); +} + +TEST_F(CommandGraphTest, LinearInOrderQueue) { + sycl::property_list Props{sycl::property::queue::in_order{}}; + queue InOrderQ{Dev, Props}; + + experimental::command_graph G{InOrderQ.get_context(), + InOrderQ.get_device()}; + G.begin_recording(InOrderQ); + BuildLinearChain(InOrderQ, /*IsInOrderQueue=*/true, /*N=*/3); + InOrderQ.submit([&](sycl::handler &cgh) { cgh.host_task([]() {}); }); + BuildLinearChain(InOrderQ, /*IsInOrderQueue=*/true, /*N=*/4); + G.end_recording(InOrderQ); + + auto Exec = G.finalize(); + auto &Impl = *getSyclObjImpl(Exec); + ValidateLinearExec(Impl, /*InOrderPartitions=*/3); +} + +TEST_F(CommandGraphTest, LinearOutOfOrderQueue) { + // Out-of-order queue but we submit a strict linear dependency chain by + // adding explicit depends_on between each node to achieve linearity. + queue OOOQ{Dev}; + experimental::command_graph G{OOOQ.get_context(), + OOOQ.get_device()}; + G.begin_recording(OOOQ); + BuildLinearChain(OOOQ, /*IsInOrderQueue=*/false, /*N=*/6); + G.end_recording(OOOQ); + + auto Exec = G.finalize(); + auto &Impl = *getSyclObjImpl(Exec); + ValidateLinearExec(Impl, /*InOrderPartitions=*/1); +} + +// Ensures non-linear graphs are creating and tracking sync points internally +// for proper scheduling and that the linear optimization is not improperly +// applied. +TEST_F(CommandGraphTest, NonLinearOutOfOrderQueue) { + queue Q{Dev}; + experimental::command_graph G{Q.get_context(), + Q.get_device()}; + G.begin_recording(Q); + // Root node + event Root = Q.submit([&](handler &h) { h.single_task([] {}); }); + // Two parallel branches depending on Root + event A = Q.submit([&](handler &h) { + h.depends_on(Root); + h.single_task([] {}); + }); + event B = Q.submit([&](handler &h) { + h.depends_on(Root); + h.single_task([] {}); + }); + // Join node depends on both A and B + Q.submit([&](handler &h) { + h.depends_on(A); + h.depends_on(B); + h.single_task([] {}); + }); + G.end_recording(Q); + + auto Exec = G.finalize(); + auto &Impl = *getSyclObjImpl(Exec); + + const int NumLinear = GraphImplTest::NumPartitionsInOrder(Impl); + const int NumSyncPoints = GraphImplTest::NumSyncPoints(Impl); + + // We should track a sync point per node for a total of 4 + EXPECT_EQ(NumSyncPoints, 4); + EXPECT_EQ(NumLinear, 0); +} From 148576d4dbcd9ba553f0066861e8b6f800e24464 Mon Sep 17 00:00:00 2001 From: Matthew Michel Date: Fri, 10 Oct 2025 14:05:16 -0700 Subject: [PATCH 2/3] Shorten alises --- sycl/unittests/Extensions/CommandGraph/Common.hpp | 5 ++--- 1 file changed, 2 insertions(+), 3 deletions(-) diff --git a/sycl/unittests/Extensions/CommandGraph/Common.hpp b/sycl/unittests/Extensions/CommandGraph/Common.hpp index 932f66c5b2283..b7b51c16baea6 100644 --- a/sycl/unittests/Extensions/CommandGraph/Common.hpp +++ b/sycl/unittests/Extensions/CommandGraph/Common.hpp @@ -30,9 +30,8 @@ using sycl::detail::getSyclObjImpl; // can access private members to analyze internal optimizations (partitions, // sync points). class GraphImplTest { - using exec_graph_impl = - sycl::ext::oneapi::experimental::detail::exec_graph_impl; - using partition = sycl::ext::oneapi::experimental::detail::partition; + using exec_graph_impl = experimental::detail::exec_graph_impl; + using partition = experimental::detail::partition; public: static int NumPartitionsInOrder(const exec_graph_impl &Impl) { From 64246b00b3f610ede83f8a45cba5402e367fb3e0 Mon Sep 17 00:00:00 2001 From: Matthew Michel Date: Fri, 10 Oct 2025 14:12:43 -0700 Subject: [PATCH 3/3] Formatting --- sycl/unittests/Extensions/CommandGraph/Common.hpp | 1 + 1 file changed, 1 insertion(+) diff --git a/sycl/unittests/Extensions/CommandGraph/Common.hpp b/sycl/unittests/Extensions/CommandGraph/Common.hpp index b7b51c16baea6..aa7c59472105e 100644 --- a/sycl/unittests/Extensions/CommandGraph/Common.hpp +++ b/sycl/unittests/Extensions/CommandGraph/Common.hpp @@ -24,6 +24,7 @@ using namespace sycl; using namespace sycl::ext::oneapi; + using sycl::detail::getSyclObjImpl; // Implement the test friend class forward declared in graph_impl.hpp so tests