From 9a8293703f5d101ac59b77ed2f751d0319fb7f0d Mon Sep 17 00:00:00 2001 From: "Kornev, Nikita" Date: Thu, 30 Jul 2026 18:20:58 +0200 Subject: [PATCH 1/2] [SYCL] Update Graph tests to use access_mode access::mode is deprecated in favor of access_mode. PR to add deprecations: https://github.com/intel/llvm/pull/22803 --- .../Inputs/buffer_copy_host2target_2d.cpp | 2 +- .../Inputs/buffer_copy_host2target_offset.cpp | 2 +- .../Graph/Inputs/buffer_copy_offsets.cpp | 4 +-- .../Graph/Inputs/buffer_copy_target2host.cpp | 2 +- .../Inputs/buffer_copy_target2host_2d.cpp | 2 +- .../Inputs/buffer_copy_target2host_offset.cpp | 2 +- .../Graph/Inputs/debug_print_graph.cpp | 2 +- .../Inputs/debug_print_graph_verbose.cpp | 2 +- .../interop-level-zero-get-native-mem.cpp | 6 ++-- sycl/test-e2e/Graph/Inputs/kernel_bundle.cpp | 2 +- .../Graph/Inputs/kernel_bundle_spirv.cpp | 8 +++--- .../Graph/Inputs/multiple_kernel_bundles.cpp | 6 ++-- .../Inputs/spec_constants_handler_api.cpp | 12 ++++---- .../spec_constants_kernel_bundle_api.cpp | 8 +++--- .../whole_update_host_task_accessor.cpp | 2 +- .../Graph/Profiling/event_profiling_info.cpp | 4 +-- .../exception_buffer_accessor.cpp | 2 +- .../Graph/Update/dyn_cgf_accessor_deps.cpp | 8 +++--- .../Graph/Update/dyn_cgf_accessor_deps2.cpp | 12 ++++---- .../Update/dyn_cgf_host_task_accessor.cpp | 4 +-- sycl/test-e2e/Graph/graph_common.hpp | 28 +++++++++---------- 21 files changed, 60 insertions(+), 60 deletions(-) diff --git a/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_2d.cpp b/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_2d.cpp index 9b865a89618c4..47d98e3cfe6f4 100644 --- a/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_2d.cpp +++ b/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_2d.cpp @@ -32,7 +32,7 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; auto NodeA = add_node(Graph, Queue, [&](handler &CGH) { - auto AccA = BufferA.get_access(CGH); + auto AccA = BufferA.get_access(CGH); CGH.copy(DataB.data(), AccA); }); diff --git a/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp b/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp index 273ff371ca58e..7853ab87415e5 100644 --- a/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp +++ b/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp @@ -30,7 +30,7 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; auto NodeA = add_node(Graph, Queue, [&](handler &CGH) { - auto AccA = BufferA.get_access(CGH, range<1>(Size), + auto AccA = BufferA.get_access(CGH, range<1>(Size), id<1>(Offset)); CGH.copy(DataB.data(), AccA); }); diff --git a/sycl/test-e2e/Graph/Inputs/buffer_copy_offsets.cpp b/sycl/test-e2e/Graph/Inputs/buffer_copy_offsets.cpp index 2e22b366cd8ba..6d595178b8122 100644 --- a/sycl/test-e2e/Graph/Inputs/buffer_copy_offsets.cpp +++ b/sycl/test-e2e/Graph/Inputs/buffer_copy_offsets.cpp @@ -38,9 +38,9 @@ int main() { // Copy from A to B auto NodeA = add_node(Graph, Queue, [&](handler &CGH) { - auto AccA = BufferA.get_access( + auto AccA = BufferA.get_access( CGH, range<1>(Size - OffsetSrc), id<1>(OffsetSrc)); - auto AccB = BufferB.get_access( + auto AccB = BufferB.get_access( CGH, range<1>(Size - OffsetDst), id<1>(OffsetDst)); CGH.copy(AccA, AccB); }); diff --git a/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host.cpp b/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host.cpp index 0eb84fe15ebfc..9e8a4d1788b3b 100644 --- a/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host.cpp +++ b/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host.cpp @@ -28,7 +28,7 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; auto NodeA = add_node(Graph, Queue, [&](handler &CGH) { - auto AccA = BufferA.get_access(CGH); + auto AccA = BufferA.get_access(CGH); CGH.copy(AccA, DataB.data()); }); diff --git a/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_2d.cpp b/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_2d.cpp index 1a5ea34c5b0e2..dad5149421c88 100644 --- a/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_2d.cpp +++ b/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_2d.cpp @@ -33,7 +33,7 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; auto NodeA = add_node(Graph, Queue, [&](handler &CGH) { - auto AccA = BufferA.get_access(CGH); + auto AccA = BufferA.get_access(CGH); CGH.copy(AccA, DataB.data()); }); diff --git a/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_offset.cpp b/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_offset.cpp index a4ccbb0f795f8..25d8a5039df32 100644 --- a/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_offset.cpp +++ b/sycl/test-e2e/Graph/Inputs/buffer_copy_target2host_offset.cpp @@ -31,7 +31,7 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; auto NodeA = add_node(Graph, Queue, [&](handler &CGH) { - auto AccA = BufferA.get_access( + auto AccA = BufferA.get_access( CGH, range<1>(Size - Offset), id<1>(Offset)); CGH.copy(AccA, DataB.data()); }); diff --git a/sycl/test-e2e/Graph/Inputs/debug_print_graph.cpp b/sycl/test-e2e/Graph/Inputs/debug_print_graph.cpp index ee2fddb8f414d..2001a546dcf69 100644 --- a/sycl/test-e2e/Graph/Inputs/debug_print_graph.cpp +++ b/sycl/test-e2e/Graph/Inputs/debug_print_graph.cpp @@ -38,7 +38,7 @@ int main() { }); auto Last = add_node(Graph, Queue, [&](handler &CGH) { - auto Acc = BufferC.get_access(CGH); + auto Acc = BufferC.get_access(CGH); CGH.copy(Acc, DataD.data()); }); diff --git a/sycl/test-e2e/Graph/Inputs/debug_print_graph_verbose.cpp b/sycl/test-e2e/Graph/Inputs/debug_print_graph_verbose.cpp index 94f1e75832138..45d70de32c505 100644 --- a/sycl/test-e2e/Graph/Inputs/debug_print_graph_verbose.cpp +++ b/sycl/test-e2e/Graph/Inputs/debug_print_graph_verbose.cpp @@ -38,7 +38,7 @@ int main() { }); auto Last = add_node(Graph, Queue, [&](handler &CGH) { - auto AccA = BufferC.get_access(CGH); + auto AccA = BufferC.get_access(CGH); CGH.copy(AccA, DataD.data()); }); diff --git a/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp b/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp index 01696a6dbe388..70e2b4a97cc66 100644 --- a/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp +++ b/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp @@ -87,7 +87,7 @@ int main() { auto NodeA = add_node(Graph, Queue, [&](sycl::handler &CGH) { auto Acc = - BufferInterop.get_access(CGH); + BufferInterop.get_access(CGH); CGH.single_task([=]() { for (int i = 0; i < 12; i++) { Acc[i] = 99; @@ -100,7 +100,7 @@ int main() { [&](sycl::handler &CGH) { depends_on_helper(CGH, NodeA); auto Acc = - BufferInterop.get_access(CGH); + BufferInterop.get_access(CGH); CGH.single_task([=]() { for (int i = 0; i < 12; i++) { Acc[i] *= 2; @@ -113,7 +113,7 @@ int main() { Graph, Queue, [&](handler &CGH) { depends_on_helper(CGH, NodeB); - auto BufferAcc = BufferInterop.get_access(CGH); + auto BufferAcc = BufferInterop.get_access(CGH); CGH.host_task([=](const interop_handle &IH) { void *DevicePtr = IH.get_native_mem(BufferAcc); diff --git a/sycl/test-e2e/Graph/Inputs/kernel_bundle.cpp b/sycl/test-e2e/Graph/Inputs/kernel_bundle.cpp index b7a15992eaafb..e5f2ddd1f68e4 100644 --- a/sycl/test-e2e/Graph/Inputs/kernel_bundle.cpp +++ b/sycl/test-e2e/Graph/Inputs/kernel_bundle.cpp @@ -31,7 +31,7 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; add_node(Graph, Queue, ([&](sycl::handler &CGH) { - auto Acc = Buf.get_access(CGH); + auto Acc = Buf.get_access(CGH); CGH.use_kernel_bundle(KernelBundleExecutable); CGH.single_task([=]() { Acc[0] = 42; }); })); diff --git a/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp b/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp index a836258176c89..a7eed51462551 100644 --- a/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp +++ b/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp @@ -41,18 +41,18 @@ int main(int, char **argv) { add_node(Graph, Queue, ([&](sycl::handler &CGH) { CGH.set_arg( - 0, input_buffer.get_access(CGH)); + 0, input_buffer.get_access(CGH)); CGH.set_arg( - 1, output_buffer.get_access(CGH)); + 1, output_buffer.get_access(CGH)); CGH.parallel_for(sycl::range<1>{N}, kernel); })); add_node(Graph, Queue, ([&](sycl::handler &CGH) { CGH.set_arg( - 0, input_buffer.get_access(CGH)); + 0, input_buffer.get_access(CGH)); CGH.set_arg( 1, - output_buffer2.get_access(CGH)); + output_buffer2.get_access(CGH)); CGH.parallel_for(sycl::range<1>{N}, kernel); })); diff --git a/sycl/test-e2e/Graph/Inputs/multiple_kernel_bundles.cpp b/sycl/test-e2e/Graph/Inputs/multiple_kernel_bundles.cpp index 5c49cf66c3b80..cd058c79d1a13 100644 --- a/sycl/test-e2e/Graph/Inputs/multiple_kernel_bundles.cpp +++ b/sycl/test-e2e/Graph/Inputs/multiple_kernel_bundles.cpp @@ -56,14 +56,14 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; add_node(Graph, Queue, ([&](sycl::handler &CGH) { - auto Acc = Buf1.get_access(CGH); + auto Acc = Buf1.get_access(CGH); CGH.use_kernel_bundle(KernelBundleExecutable1); CGH.single_task([=]() { Acc[0] = 42; }); })); add_node(Graph, Queue, ([&](handler &CGH) { auto DataA = - BufferA.template get_access(CGH); + BufferA.template get_access(CGH); CGH.use_kernel_bundle(KernelBundleExecutable1); CGH.parallel_for(range<1>{Size}, [=](item<1> Id) { DataA[Id]++; }); @@ -77,7 +77,7 @@ int main() { std::error_code ExceptionCode = make_error_code(sycl::errc::success); try { Graph.add([&](sycl::handler &CGH) { - auto Acc = Buf2.get_access(CGH); + auto Acc = Buf2.get_access(CGH); CGH.use_kernel_bundle(KernelBundleExecutable2); CGH.single_task([=]() { Acc[0] = 24; }); }); diff --git a/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp b/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp index eb2ac17eb72ba..98d3e730f4b5b 100644 --- a/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp +++ b/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp @@ -74,11 +74,11 @@ bool test_default_values(sycl::queue Queue) { add_node(Graph, Queue, ([&](sycl::handler &CGH) { auto IntAcc = - IntBuffer.get_access(CGH); + IntBuffer.get_access(CGH); auto IntAcc2 = - IntBuffer2.get_access(CGH); + IntBuffer2.get_access(CGH); auto FloatAcc = - FloatBuffer.get_access(CGH); + FloatBuffer.get_access(CGH); CGH.single_task( [=](sycl::kernel_handler KH) { @@ -173,10 +173,10 @@ bool test_set_and_get_on_device(sycl::queue Queue) { add_node( Graph, Queue, ([&](sycl::handler &CGH) { - auto IntAcc = IntBuffer.get_access(CGH); - auto IntAcc2 = IntBuffer2.get_access(CGH); + auto IntAcc = IntBuffer.get_access(CGH); + auto IntAcc2 = IntBuffer2.get_access(CGH); auto FloatAcc = - FloatBuffer.get_access(CGH); + FloatBuffer.get_access(CGH); CGH.set_specialization_constant(NewIntValue); CGH.set_specialization_constant(NewIntValue2); diff --git a/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp b/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp index 7604b5bdfd5db..f8683dc8e247e 100644 --- a/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp +++ b/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp @@ -82,9 +82,9 @@ bool test_default_values(sycl::queue Queue) { add_node(Graph, Queue, ([&](sycl::handler &CGH) { CGH.use_kernel_bundle(ExecBundle); auto IntAcc = - IntBuffer.get_access(CGH); + IntBuffer.get_access(CGH); auto FloatAcc = - FloatBuffer.get_access(CGH); + FloatBuffer.get_access(CGH); CGH.single_task( [=](sycl::kernel_handler KH) { @@ -205,9 +205,9 @@ bool test_set_and_get_on_device(sycl::queue Queue) { add_node( Graph, Queue, ([&](sycl::handler &CGH) { CGH.use_kernel_bundle(ExecBundle); - auto IntAcc = IntBuffer.get_access(CGH); + auto IntAcc = IntBuffer.get_access(CGH); auto FloatAcc = - FloatBuffer.get_access(CGH); + FloatBuffer.get_access(CGH); CGH.single_task([=](sycl::kernel_handler KH) { IntAcc[0] = KH.get_specialization_constant(); diff --git a/sycl/test-e2e/Graph/Inputs/whole_update_host_task_accessor.cpp b/sycl/test-e2e/Graph/Inputs/whole_update_host_task_accessor.cpp index d94e8e3fe2284..36be6f8a2fe0b 100644 --- a/sycl/test-e2e/Graph/Inputs/whole_update_host_task_accessor.cpp +++ b/sycl/test-e2e/Graph/Inputs/whole_update_host_task_accessor.cpp @@ -32,7 +32,7 @@ void add_nodes_to_graph( auto HostTaskOp = add_node( Graph, Queue, [&](handler &CGH) { - auto AccC = BufferC.get_access(CGH); depends_on_helper(CGH, LastOperation); CGH.host_task([=]() { diff --git a/sycl/test-e2e/Graph/Profiling/event_profiling_info.cpp b/sycl/test-e2e/Graph/Profiling/event_profiling_info.cpp index 9531094bc0578..d176760678a0a 100644 --- a/sycl/test-e2e/Graph/Profiling/event_profiling_info.cpp +++ b/sycl/test-e2e/Graph/Profiling/event_profiling_info.cpp @@ -48,9 +48,9 @@ int main() { CopyGraph.begin_recording(Queue); Queue.submit([&](sycl::handler &Cgh) { - accessor AccessorFrom( + accessor AccessorFrom( BufferFrom, Cgh, range<1>(Size)); - accessor AccessorTo( + accessor AccessorTo( BufferTo, Cgh, range<1>(Size)); Cgh.copy(AccessorFrom, AccessorTo); }); diff --git a/sycl/test-e2e/Graph/RecordReplay/NativeRecording/exception_buffer_accessor.cpp b/sycl/test-e2e/Graph/RecordReplay/NativeRecording/exception_buffer_accessor.cpp index 1e253ca136f89..625f13a6d40ea 100644 --- a/sycl/test-e2e/Graph/RecordReplay/NativeRecording/exception_buffer_accessor.cpp +++ b/sycl/test-e2e/Graph/RecordReplay/NativeRecording/exception_buffer_accessor.cpp @@ -30,7 +30,7 @@ int main() { if (!expectException( [&]() { Queue.submit([&](handler &CGH) { - auto Acc = Buf.get_access(CGH); + auto Acc = Buf.get_access(CGH); CGH.parallel_for(sycl::range<1>{N}, [=](sycl::id<1> Idx) { Acc[Idx] = 0; }); }); diff --git a/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps.cpp b/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps.cpp index 4cdf4550a1835..7883072a922bd 100644 --- a/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps.cpp +++ b/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps.cpp @@ -24,20 +24,20 @@ int main() { {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; auto RootNode = Graph.add([&](handler &CGH) { - auto Acc = Buf.get_access(CGH); + auto Acc = Buf.get_access(CGH); CGH.parallel_for(Size, [=](item<1> Item) { Acc[Item.get_id()] = 1; }); }); int PatternA = 42; auto CGFA = [&](handler &CGH) { - auto Acc = Buf.get_access(CGH); + auto Acc = Buf.get_access(CGH); CGH.parallel_for(Size, [=](item<1> Item) { Acc[Item.get_id()] += PatternA; }); }; int PatternB = 0xA; auto CGFB = [&](handler &CGH) { - auto Acc = Buf.get_access(CGH); + auto Acc = Buf.get_access(CGH); CGH.parallel_for(Size, [=](item<1> Item) { Acc[Item.get_id()] += PatternB; }); }; @@ -46,7 +46,7 @@ int main() { auto DynamicCGNode = Graph.add(DynamicCG); auto LeafNode = Graph.add([&](handler &CGH) { - auto Acc = Buf.get_access(CGH); + auto Acc = Buf.get_access(CGH); CGH.parallel_for( Size, [=](item<1> Item) { Ptr[Item.get_id()] = Acc[Item.get_id()]; }); }); diff --git a/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps2.cpp b/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps2.cpp index 870b67bad722f..b69ad90b4d173 100644 --- a/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps2.cpp +++ b/sycl/test-e2e/Graph/Update/dyn_cgf_accessor_deps2.cpp @@ -29,8 +29,8 @@ int main() { int InitA = 4; int InitB = -4; auto RootNode = Graph.add([&](handler &CGH) { - auto AccA = BufA.get_access(CGH); - auto AccB = BufB.get_access(CGH); + auto AccA = BufA.get_access(CGH); + auto AccB = BufB.get_access(CGH); CGH.parallel_for(Size, [=](item<1> Item) { AccA[Item.get_id()] = InitA; AccB[Item.get_id()] = InitB; @@ -39,14 +39,14 @@ int main() { int PatternA = 42; auto CGFA = [&](handler &CGH) { - auto AccA = BufA.get_access(CGH); + auto AccA = BufA.get_access(CGH); CGH.parallel_for(Size, [=](item<1> Item) { AccA[Item.get_id()] += PatternA; }); }; int PatternB = 0xA; auto CGFB = [&](handler &CGH) { - auto AccB = BufB.get_access(CGH); + auto AccB = BufB.get_access(CGH); CGH.parallel_for(Size, [=](item<1> Item) { AccB[Item.get_id()] += PatternB; }); }; @@ -55,8 +55,8 @@ int main() { auto DynamicCGNode = Graph.add(DynamicCG); auto LeafNode = Graph.add([&](handler &CGH) { - auto AccA = BufA.get_access(CGH); - auto AccB = BufB.get_access(CGH); + auto AccA = BufA.get_access(CGH); + auto AccB = BufB.get_access(CGH); CGH.parallel_for(Size, [=](item<1> Item) { Ptr[Item.get_id()] = AccA[Item.get_id()] + AccB[Item.get_id()]; }); diff --git a/sycl/test-e2e/Graph/Update/dyn_cgf_host_task_accessor.cpp b/sycl/test-e2e/Graph/Update/dyn_cgf_host_task_accessor.cpp index 68fce4042a043..6f7a8c76c268b 100644 --- a/sycl/test-e2e/Graph/Update/dyn_cgf_host_task_accessor.cpp +++ b/sycl/test-e2e/Graph/Update/dyn_cgf_host_task_accessor.cpp @@ -71,7 +71,7 @@ int main() { // Create two different command groups for the host task auto CGFA = [&](handler &CGH) { - auto AccC = BufferC.get_access(CGH); CGH.host_task([=]() { for (size_t i = 0; i < Size; i++) { @@ -80,7 +80,7 @@ int main() { }); }; auto CGFB = [&](handler &CGH) { - auto AccC = BufferC.get_access(CGH); CGH.host_task([=]() { for (size_t i = 0; i < Size; i++) { diff --git a/sycl/test-e2e/Graph/graph_common.hpp b/sycl/test-e2e/Graph/graph_common.hpp index 9d7bbbe1c044d..3ab894f6ac934 100644 --- a/sycl/test-e2e/Graph/graph_common.hpp +++ b/sycl/test-e2e/Graph/graph_common.hpp @@ -100,15 +100,15 @@ event run_kernels(queue Q, const size_t Size, buffer BufferA, buffer BufferB, buffer BufferC) { // Read & write Buffer A. Q.submit([&](handler &CGH) { - auto DataA = BufferA.template get_access(CGH); + auto DataA = BufferA.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataA[Id]++; }); }); // Reads Buffer A. // Read & Write Buffer B. Q.submit([&](handler &CGH) { - auto DataA = BufferA.template get_access(CGH); - auto DataB = BufferB.template get_access(CGH); + auto DataA = BufferA.template get_access(CGH); + auto DataB = BufferB.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataB[Id] += DataA[Id]; }); }); @@ -116,16 +116,16 @@ event run_kernels(queue Q, const size_t Size, buffer BufferA, // Reads Buffer A. // Read & writes Buffer C Q.submit([&](handler &CGH) { - auto DataA = BufferA.template get_access(CGH); - auto DataC = BufferC.template get_access(CGH); + auto DataA = BufferA.template get_access(CGH); + auto DataC = BufferC.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataC[Id] -= DataA[Id]; }); }); // Read & write Buffers B and C. auto ExitEvent = Q.submit([&](handler &CGH) { - auto DataB = BufferB.template get_access(CGH); - auto DataC = BufferC.template get_access(CGH); + auto DataB = BufferB.template get_access(CGH); + auto DataC = BufferC.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataB[Id]--; DataC[Id]--; @@ -151,15 +151,15 @@ add_kernels(exp_ext::command_graph Graph, buffer BufferC) { // Read & write Buffer A Graph.add([&](handler &CGH) { - auto DataA = BufferA.template get_access(CGH); + auto DataA = BufferA.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataA[Id]++; }); }); // Reads Buffer A // Read & Write Buffer B Graph.add([&](handler &CGH) { - auto DataA = BufferA.template get_access(CGH); - auto DataB = BufferB.template get_access(CGH); + auto DataA = BufferA.template get_access(CGH); + auto DataB = BufferB.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataB[Id] += DataA[Id]; }); }); @@ -167,16 +167,16 @@ add_kernels(exp_ext::command_graph Graph, // Reads Buffer A // Read & writes Buffer C Graph.add([&](handler &CGH) { - auto DataA = BufferA.template get_access(CGH); - auto DataC = BufferC.template get_access(CGH); + auto DataA = BufferA.template get_access(CGH); + auto DataC = BufferC.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataC[Id] -= DataA[Id]; }); }); // Read & write Buffers B and C auto ExitNode = Graph.add([&](handler &CGH) { - auto DataB = BufferB.template get_access(CGH); - auto DataC = BufferC.template get_access(CGH); + auto DataB = BufferB.template get_access(CGH); + auto DataC = BufferC.template get_access(CGH); CGH.parallel_for(range<1>(Size), [=](item<1> Id) { DataB[Id]--; DataC[Id]--; From aa91fdadd0150887408905f2fe02d5140af26689 Mon Sep 17 00:00:00 2001 From: "Kornev, Nikita" Date: Thu, 30 Jul 2026 18:59:29 +0200 Subject: [PATCH 2/2] format --- .../Inputs/buffer_copy_host2target_offset.cpp | 2 +- .../interop-level-zero-get-native-mem.cpp | 3 +- .../Graph/Inputs/kernel_bundle_spirv.cpp | 31 +++++++++---------- .../Inputs/spec_constants_handler_api.cpp | 3 +- .../spec_constants_kernel_bundle_api.cpp | 3 +- 5 files changed, 19 insertions(+), 23 deletions(-) diff --git a/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp b/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp index 7853ab87415e5..6ed896e4c8118 100644 --- a/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp +++ b/sycl/test-e2e/Graph/Inputs/buffer_copy_host2target_offset.cpp @@ -31,7 +31,7 @@ int main() { auto NodeA = add_node(Graph, Queue, [&](handler &CGH) { auto AccA = BufferA.get_access(CGH, range<1>(Size), - id<1>(Offset)); + id<1>(Offset)); CGH.copy(DataB.data(), AccA); }); diff --git a/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp b/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp index 70e2b4a97cc66..d1139b7ad6abf 100644 --- a/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp +++ b/sycl/test-e2e/Graph/Inputs/interop-level-zero-get-native-mem.cpp @@ -86,8 +86,7 @@ int main() { BufferInteropInput, Context); auto NodeA = add_node(Graph, Queue, [&](sycl::handler &CGH) { - auto Acc = - BufferInterop.get_access(CGH); + auto Acc = BufferInterop.get_access(CGH); CGH.single_task([=]() { for (int i = 0; i < 12; i++) { Acc[i] = 99; diff --git a/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp b/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp index a7eed51462551..904247c678baa 100644 --- a/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp +++ b/sycl/test-e2e/Graph/Inputs/kernel_bundle_spirv.cpp @@ -39,22 +39,21 @@ int main(int, char **argv) { Queue.get_device(), {exp_ext::property::graph::assume_buffer_outlives_graph{}}}; - add_node(Graph, Queue, ([&](sycl::handler &CGH) { - CGH.set_arg( - 0, input_buffer.get_access(CGH)); - CGH.set_arg( - 1, output_buffer.get_access(CGH)); - CGH.parallel_for(sycl::range<1>{N}, kernel); - })); - - add_node(Graph, Queue, ([&](sycl::handler &CGH) { - CGH.set_arg( - 0, input_buffer.get_access(CGH)); - CGH.set_arg( - 1, - output_buffer2.get_access(CGH)); - CGH.parallel_for(sycl::range<1>{N}, kernel); - })); + add_node( + Graph, Queue, ([&](sycl::handler &CGH) { + CGH.set_arg(0, input_buffer.get_access(CGH)); + CGH.set_arg(1, + output_buffer.get_access(CGH)); + CGH.parallel_for(sycl::range<1>{N}, kernel); + })); + + add_node( + Graph, Queue, ([&](sycl::handler &CGH) { + CGH.set_arg(0, input_buffer.get_access(CGH)); + CGH.set_arg(1, + output_buffer2.get_access(CGH)); + CGH.parallel_for(sycl::range<1>{N}, kernel); + })); auto GraphExec = Graph.finalize(); diff --git a/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp b/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp index 98d3e730f4b5b..8d7eec4c55905 100644 --- a/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp +++ b/sycl/test-e2e/Graph/Inputs/spec_constants_handler_api.cpp @@ -175,8 +175,7 @@ bool test_set_and_get_on_device(sycl::queue Queue) { Graph, Queue, ([&](sycl::handler &CGH) { auto IntAcc = IntBuffer.get_access(CGH); auto IntAcc2 = IntBuffer2.get_access(CGH); - auto FloatAcc = - FloatBuffer.get_access(CGH); + auto FloatAcc = FloatBuffer.get_access(CGH); CGH.set_specialization_constant(NewIntValue); CGH.set_specialization_constant(NewIntValue2); diff --git a/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp b/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp index f8683dc8e247e..7c92fcacc6ca0 100644 --- a/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp +++ b/sycl/test-e2e/Graph/Inputs/spec_constants_kernel_bundle_api.cpp @@ -206,8 +206,7 @@ bool test_set_and_get_on_device(sycl::queue Queue) { Graph, Queue, ([&](sycl::handler &CGH) { CGH.use_kernel_bundle(ExecBundle); auto IntAcc = IntBuffer.get_access(CGH); - auto FloatAcc = - FloatBuffer.get_access(CGH); + auto FloatAcc = FloatBuffer.get_access(CGH); CGH.single_task([=](sycl::kernel_handler KH) { IntAcc[0] = KH.get_specialization_constant();