From caa70586c85d40f524500100c4550a84b5e4e80c Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Thu, 21 Dec 2023 16:21:41 +0000
Subject: [PATCH 01/12] [SYCL][Graph] HIP Support testing

Co-authored-by: Andrey Alekseenko <andrey.alekseenko@scilifelab.se>
---
 sycl/plugins/unified_runtime/CMakeLists.txt | 11 +++--------
 1 file changed, 3 insertions(+), 8 deletions(-)

diff --git a/sycl/plugins/unified_runtime/CMakeLists.txt b/sycl/plugins/unified_runtime/CMakeLists.txt
index 1ce7db18e0025..7a0b2042521a4 100644
--- a/sycl/plugins/unified_runtime/CMakeLists.txt
+++ b/sycl/plugins/unified_runtime/CMakeLists.txt
@@ -56,14 +56,9 @@ endif()
 if(SYCL_PI_UR_USE_FETCH_CONTENT)
   include(FetchContent)
 
-  set(UNIFIED_RUNTIME_REPO "https://github.com/oneapi-src/unified-runtime.git")
-  # commit 8d1486a607666763413fea3c70b74a3dd8f73c0d
-  # Author: aarongreig <aaron.greig@codeplay.com>
-  # Date: Mon Dec 18 13:15:43 2023 +0000
-  #     Merge pull request #939 from steffenlarsen/steffen/virtual_mem_adapters
-  #     [UR][CUDA][L0][HIP] Add virtual memory adapter implementations
-  set(UNIFIED_RUNTIME_TAG 8d1486a607666763413fea3c70b74a3dd8f73c0d)
-
+  set(UNIFIED_RUNTIME_REPO "https://github.com/bensuo/unified-runtime.git")
+  set(UNIFIED_RUNTIME_TAG cmdbuf-hip-support)
+  
   if(SYCL_PI_UR_OVERRIDE_FETCH_CONTENT_REPO)
     set(UNIFIED_RUNTIME_REPO "${SYCL_PI_UR_OVERRIDE_FETCH_CONTENT_REPO}")
   endif()

From 816cfd3986bd98482c21fda8d011f21c7c24b4bc Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Thu, 21 Dec 2023 15:08:53 +0000
Subject: [PATCH 02/12] [SYCL][Graph] Fix OpenCL backend test fails with CPU
 device

Adds missing variable initializations.
Temporary disable `work_group_size_prop.cpp` because the failure results from a bug in the emulation layer.
---
 sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp     | 8 ++++++--
 sycl/test-e2e/Graph/RecordReplay/dotp_in_order.cpp        | 1 +
 .../Graph/RecordReplay/dotp_in_order_with_empty_nodes.cpp | 1 +
 sycl/test-e2e/Graph/RecordReplay/dotp_multiple_queues.cpp | 1 +
 sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp | 8 ++++++--
 5 files changed, 15 insertions(+), 4 deletions(-)

diff --git a/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp b/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
index efc03f077856e..60cdb64545aea 100644
--- a/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
+++ b/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
@@ -5,8 +5,12 @@
 //
 // CHECK-NOT: LEAK
 
-// Temporarily disabled for CUDA.
-// XFAIL: cuda
+// Temporarily disabled for CUDA and OpenCL
+// The OpenCL emulation layer does not return `CL_INVALID_WORK_GROUP_SIZE` as it
+// should. So the Sycl graph support cannot correctly catch the error and throw
+// the approriate exception for negative test. An issue has been reported
+// https://github.com/bashbaug/SimpleOpenCLSamples/issues/95
+// XFAIL: cuda, opencl
 // Note: failing negative test with HIP in the original test
 // TODO: disable hip when HIP backend will be supported by Graph
 
diff --git a/sycl/test-e2e/Graph/RecordReplay/dotp_in_order.cpp b/sycl/test-e2e/Graph/RecordReplay/dotp_in_order.cpp
index 9e7bd6ff6125d..6797d32d18290 100644
--- a/sycl/test-e2e/Graph/RecordReplay/dotp_in_order.cpp
+++ b/sycl/test-e2e/Graph/RecordReplay/dotp_in_order.cpp
@@ -22,6 +22,7 @@ int main() {
   exp_ext::command_graph Graph{Queue.get_context(), Queue.get_device()};
 
   int *Dotp = malloc_device<int>(1, Queue);
+  Queue.memset(Dotp, 0, sizeof(int)).wait();
 
   const size_t N = 10;
   int *X = malloc_device<int>(N, Queue);
diff --git a/sycl/test-e2e/Graph/RecordReplay/dotp_in_order_with_empty_nodes.cpp b/sycl/test-e2e/Graph/RecordReplay/dotp_in_order_with_empty_nodes.cpp
index 9abf37bdaf01b..01ff2476b845a 100644
--- a/sycl/test-e2e/Graph/RecordReplay/dotp_in_order_with_empty_nodes.cpp
+++ b/sycl/test-e2e/Graph/RecordReplay/dotp_in_order_with_empty_nodes.cpp
@@ -24,6 +24,7 @@ int main() {
   exp_ext::command_graph Graph{Queue.get_context(), Queue.get_device()};
 
   int *Dotp = malloc_device<int>(1, Queue);
+  Queue.memset(Dotp, 0, sizeof(int)).wait();
 
   const size_t N = 10;
   int *X = malloc_device<int>(N, Queue);
diff --git a/sycl/test-e2e/Graph/RecordReplay/dotp_multiple_queues.cpp b/sycl/test-e2e/Graph/RecordReplay/dotp_multiple_queues.cpp
index c9bc113cc9abc..f518f5f0e4b53 100644
--- a/sycl/test-e2e/Graph/RecordReplay/dotp_multiple_queues.cpp
+++ b/sycl/test-e2e/Graph/RecordReplay/dotp_multiple_queues.cpp
@@ -24,6 +24,7 @@ int main() {
   exp_ext::command_graph Graph{QueueA.get_context(), QueueA.get_device()};
 
   int *Dotp = malloc_device<int>(1, QueueA);
+  QueueA.memset(Dotp, 0, sizeof(int)).wait();
 
   const size_t N = 10;
   int *X = malloc_device<int>(N, QueueA);
diff --git a/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp b/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
index 40c7f1d16e2fc..b0e9121806fa6 100644
--- a/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
+++ b/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
@@ -5,8 +5,12 @@
 //
 // CHECK-NOT: LEAK
 
-// Temporarily disabled for CUDA.
-// XFAIL: cuda
+// Temporarily disabled for CUDA and OpenCL
+// The OpenCL emulation layer does not return `CL_INVALID_WORK_GROUP_SIZE` as it
+// should. So the Sycl graph support cannot correctly catch the error and throw
+// the approriate exception for negative test. An issue has been reported
+// https://github.com/bashbaug/SimpleOpenCLSamples/issues/95
+// XFAIL: cuda, opencl
 // Note: failing negative test with HIP in the original test
 // TODO: disable hip when HIP backend will be supported by Graph
 

From 9ed12b6ef70a4042195c6f995a125fee6afba700 Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Fri, 22 Dec 2023 11:10:11 +0000
Subject: [PATCH 03/12] Disables unsupported test for hip + adds hip as
 supported in the test that checks backend support

---
 sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp     | 3 +--
 sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp | 3 +--
 sycl/test-e2e/Graph/device_query.cpp                      | 3 ++-
 3 files changed, 4 insertions(+), 5 deletions(-)

diff --git a/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp b/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
index 60cdb64545aea..ab28e4a90e1e6 100644
--- a/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
+++ b/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
@@ -10,9 +10,8 @@
 // should. So the Sycl graph support cannot correctly catch the error and throw
 // the approriate exception for negative test. An issue has been reported
 // https://github.com/bashbaug/SimpleOpenCLSamples/issues/95
-// XFAIL: cuda, opencl
+// XFAIL: cuda, opencl, hip
 // Note: failing negative test with HIP in the original test
-// TODO: disable hip when HIP backend will be supported by Graph
 
 #define GRAPH_E2E_EXPLICIT
 
diff --git a/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp b/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
index b0e9121806fa6..82c631a43f254 100644
--- a/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
+++ b/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
@@ -10,9 +10,8 @@
 // should. So the Sycl graph support cannot correctly catch the error and throw
 // the approriate exception for negative test. An issue has been reported
 // https://github.com/bashbaug/SimpleOpenCLSamples/issues/95
-// XFAIL: cuda, opencl
+// XFAIL: cuda, opencl, hip
 // Note: failing negative test with HIP in the original test
-// TODO: disable hip when HIP backend will be supported by Graph
 
 #define GRAPH_E2E_RECORD_REPLAY
 
diff --git a/sycl/test-e2e/Graph/device_query.cpp b/sycl/test-e2e/Graph/device_query.cpp
index f1f2e680c323a..68dc85fce9332 100644
--- a/sycl/test-e2e/Graph/device_query.cpp
+++ b/sycl/test-e2e/Graph/device_query.cpp
@@ -20,7 +20,8 @@ int main() {
   auto Backend = Device.get_backend();
 
   if ((Backend == backend::ext_oneapi_level_zero) ||
-      (Backend == backend::ext_oneapi_cuda)) {
+      (Backend == backend::ext_oneapi_cuda) ||
+      (Backend == backend::hip) {
     assert(SupportsGraphs == exp_ext::graph_support_level::native);
   } else if (Backend == backend::opencl) {
     // OpenCL backend support is conditional on the cl_khr_command_buffer

From 8a15fa62daabc9b487d58b736fc202443cd073d5 Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Fri, 22 Dec 2023 14:31:27 +0000
Subject: [PATCH 04/12] Fix hip backend name

---
 sycl/test-e2e/Graph/device_query.cpp | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/sycl/test-e2e/Graph/device_query.cpp b/sycl/test-e2e/Graph/device_query.cpp
index 68dc85fce9332..1d740242c8b2b 100644
--- a/sycl/test-e2e/Graph/device_query.cpp
+++ b/sycl/test-e2e/Graph/device_query.cpp
@@ -21,7 +21,7 @@ int main() {
 
   if ((Backend == backend::ext_oneapi_level_zero) ||
       (Backend == backend::ext_oneapi_cuda) ||
-      (Backend == backend::hip) {
+      (Backend == backend::ext_oneapi_hip) {
     assert(SupportsGraphs == exp_ext::graph_support_level::native);
   } else if (Backend == backend::opencl) {
     // OpenCL backend support is conditional on the cl_khr_command_buffer

From 4954403cf329283dbb52af52e8ea4a6ade06ed6a Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Fri, 22 Dec 2023 15:09:06 +0000
Subject: [PATCH 05/12] Typo

---
 sycl/test-e2e/Graph/device_query.cpp | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/sycl/test-e2e/Graph/device_query.cpp b/sycl/test-e2e/Graph/device_query.cpp
index 1d740242c8b2b..3513aa74fa5b7 100644
--- a/sycl/test-e2e/Graph/device_query.cpp
+++ b/sycl/test-e2e/Graph/device_query.cpp
@@ -21,7 +21,7 @@ int main() {
 
   if ((Backend == backend::ext_oneapi_level_zero) ||
       (Backend == backend::ext_oneapi_cuda) ||
-      (Backend == backend::ext_oneapi_hip) {
+      (Backend == backend::ext_oneapi_hip)) {
     assert(SupportsGraphs == exp_ext::graph_support_level::native);
   } else if (Backend == backend::opencl) {
     // OpenCL backend support is conditional on the cl_khr_command_buffer

From 68bf06e5084b9a15e7281adb3d07ae5fe2dd1d23 Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Tue, 2 Jan 2024 11:12:37 +0000
Subject: [PATCH 06/12] Updates design doc

---
 sycl/doc/design/CommandGraph.md               | 23 ++++++++++++++++++-
 .../design/images/SYCL-Graph-Architecture.svg |  2 +-
 2 files changed, 23 insertions(+), 2 deletions(-)

diff --git a/sycl/doc/design/CommandGraph.md b/sycl/doc/design/CommandGraph.md
index 7c405d5ca791b..a2e635a5a95b0 100644
--- a/sycl/doc/design/CommandGraph.md
+++ b/sycl/doc/design/CommandGraph.md
@@ -208,7 +208,7 @@ Implementation of UR command-buffers
 for each of the supported SYCL 2020 backends.
 
 Backends which are implemented currently are: [Level Zero](#level-zero),
-[CUDA](#cuda), and partial support for [OpenCL](#opencl).
+[CUDA](#cuda), [HIP](#hip) and partial support for [OpenCL](#opencl).
 
 ### Level Zero
 
@@ -312,6 +312,27 @@ An executable CUDA Graph, which contains all commands and synchronization
 information, is saved in the UR command-buffer to allow for efficient
 graph resubmission.
 
+### HIP
+
+The HIP backend offers a Graph managemenet API very similar to CUDA Graph
+feature for batching series of operations.
+The SYCL Graph HIP backend implementation is therefore very similar to that of CUDA.
+
+UR commands (e.g. kernels) are mapped as graph nodes using the
+[HIP Management API](https://docs.amd.com/projects/HIP/en/docs-5.5.0/doxygen/html/group___graph.html).
+Synchronization between commands (UR sync-points) is implemented
+using graph dependencies.
+Executable HIP Graphs can be submitted to a HIP stream
+in the same way as regular kernels.
+The HIP backend enables enqueuing events to wait for into a stream.
+It also allows signaling the completion of a submission with an event.
+Therefore, submitting a UR command-buffer consists only of submitting to a stream
+the executable HIP Graph that represent this series of operations.
+
+An executable HIP Graph, which contains all commands and synchronization
+information, is saved in the UR command-buffer to allow for efficient
+graph resubmission.
+
 ### OpenCL
 
 SYCL-Graph is only enabled for an OpenCL backend when the
diff --git a/sycl/doc/design/images/SYCL-Graph-Architecture.svg b/sycl/doc/design/images/SYCL-Graph-Architecture.svg
index c554391f6544d..ffd3035495071 100644
--- a/sycl/doc/design/images/SYCL-Graph-Architecture.svg
+++ b/sycl/doc/design/images/SYCL-Graph-Architecture.svg
@@ -1,4 +1,4 @@
 <?xml version="1.0" encoding="UTF-8"?>
 <!-- Do not edit this file with editors other than draw.io -->
 <!DOCTYPE svg PUBLIC "-//W3C//DTD SVG 1.1//EN" "http://www.w3.org/Graphics/SVG/1.1/DTD/svg11.dtd">
-<svg xmlns="http://www.w3.org/2000/svg" style="background-color: rgb(255, 255, 255);" xmlns:xlink="http://www.w3.org/1999/xlink" version="1.1" width="661px" height="476px" viewBox="-0.5 -0.5 661 476" content="&lt;mxfile host=&quot;app.diagrams.net&quot; modified=&quot;2023-11-14T10:09:55.262Z&quot; agent=&quot;Mozilla/5.0 (Windows NT 10.0; Win64; x64; rv:109.0) Gecko/20100101 Firefox/119.0&quot; etag=&quot;SEA3Q80VPm8SFBjLcNtd&quot; version=&quot;22.1.0&quot; type=&quot;device&quot;&gt;&lt;diagram name=&quot;Page-1&quot; id=&quot;rdOIFfZhXihIw1UrQana&quot;&gt;7Vtde6I4FP41vbQP38Kltda609l1t9vddm76IARli4SJ8Wt+/SSQiJC0Uouonb1pySEJ5D3vOZxzEi/07nTVR24y+Qp9EF1oir+60K8vNK3tGOQvFawzgaEywRiFfiZSc8F9+AMwocKk89AHs0JHDGGEw6Qo9GAcAw8XZC5CcFnsFsCo+NTEHQNBcO+5kSj9N/TxJJPappLLb0E4nvAnqwq7M3V5ZyaYTVwfLrdEeu9C7yIIcXY1XXVBRLHjuGTjbl65u3kxBGJcZcD85Wa1ai8fez9s0wFOy+iN+y06gE6zcKM5WzF7W7zmEACfIMKaEOEJHMPYjXq59ArBeewD+hyFtPI+dxAmRKgS4X8A4zVTrzvHkIgmeBqxu+Ja2PJmcI488NYCGCdcNAb4jX561o+uZesBDKk+gFOA0Zp0QCBycbgoat9lJBpv+uU4kwsG9XtgF1DvJEkUeuTRMBYUUIR3OQkxuE/cFJYlMboilEEYRV0YQZSO1X0T2L5B5DOM4AvYumNrI92y3gJ/ARAGqzfh4nctRnJm5bqTNZe5yRgG6zLZMhfjUADrAsD3T927VuqjiLy3wiCepVgrneGgVsSBSjBvyxB3rLbuHghxlbuZnZArB4LckEJOJH/NYxxOwScAWVOODbLqHMM9EwTR+pGOvzR584lNlzauV4XWmrX2d+tmRbdu1e3W06EdhNz1VocEhjGebc08pIKcJxtecPdnlj7F7+tPLrI3yHmyWcr+1DEF+3yIwyAkvNgyUfIeRGtKF06nbkzvXM2DAKCiyzx/O7aPbseiNs4r7rIqGqh9UnGXJaDefbju1BtwucAOPBmjLc8Go+BAAZclMtpuktC2gOzv/wyuBwRbpT98+AQIb/zDsRBWRfKel8twKroMnvqfiM9wBNjvwCItcnwDCJ4hszXTOS3fwRW+BfEgxinEXeo7utyJ0IubYb+TXV1eXp4l+sZOv+KoTaLfPnO/ovK64E7HclpFIP7eW8DfDoa1UjqwPeBJKT2yTcN8M0erTmn91IIRVYxGzozSWlVKG6dFaU0A/o8ExGkZyIueXybo2ctSy+cRSyzrZHsQaHK2+9bIMmtKJg1L3cl2lRdAm6G7WO3sfL0+QOR9HHdy/MhbLG02GZc0RGuHFvcKuEuq+I3iroveRHTjsd+h23yk5UXubBZ6RSyLwINViLfqmKT1xPrR67yKSRu8iNl88VOv6Pyr+v4t/ZkS/XHZ+2qkQlFTd14psvEpsnWzUdrr1dGNvfOJ9NJEGTDCRHUVSvUqAfEnpJ1RkXbm/7Q7BO0MMQuuj3ZGgXiXiraDfKQxBCgkawLow8Q6NcKUyh/lTejKhCnv4lgNE0YM+2ojjE5SqJwwyqVjv58wp7tr6JwYIT+LB6tSod6TkLahVf90HsJ77U7FtZNi1U7vVJlVZmkitWFWycIxKyJoXfnhgr5HFI7j9Ib1fU5P0115RMm0tsDb5GrM/qfjRmVBANMV5zzlI+mN1iwt8NCtLMKgVTasNHPhdFEHeTTB8/AcAf4EsvTsIcUHE7HwMkSWLoxLSxZE8jlctJliXhjDGJSSSCaakYwzjMdEYOatv9NiVkt7LS+FJIUMotRgJ6Hvg1gwUo4/Q72e3bd26XgTs8LtDXtdYjzaoXJRUxadZfpKKnJQQq8ojEGLvz8lGP0UmFKGbbiLJHxJXmXLuwoPh9CkVfooSYo5qqyqsDkjU78qxbq7ANsuoGZJdtg4CFcU3OOcurRKnllyqkVr1EYq1Gs+CGwz54XKwKrqsZGtEOp/lLKN7H4KyErcQbPIirXdupFtpmouIOuI1dtmkRWPuBUOuSt37lqyAbR/eMO/XhEI6AwUOvKwqMPEUxK9pPt6MnUWP5N1qMPe6Zs3XqUZdYjJ2WCaRGBKv/U+uXHlei+Anvn8NRQi8zxqsyGlmNjsODH/WVUh+bw2bBzigYGbeZa+5Xah3M+TBCL8y6hF8gVRZaekD6cW2dm78WGc1Cbn+YASaIbPz5SYdWikVBmUuCxLoo9y2W8fffTDP6fP31++3AbD/m+LL8Hg22Nb8quq2mOlqnvdAogSqKvHSlZzsZIUWDEIHbqEhW50Zv7nY1qx1Z1akR6s2UMtpJn/8jarVOY/X9Z7PwE=&lt;/diagram&gt;&lt;/mxfile&gt;"><defs/><g><path d="M 220 75 L 220 108.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 220 113.88 L 216.5 106.88 L 220 108.63 L 223.5 106.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="0" y="34" width="440" height="41" fill="#d5e8d4" stroke="#82b366" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 55px; margin-left: 1px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Application</div></div></div></foreignObject><text x="220" y="58" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Application</text></switch></g><rect x="0" y="115" width="440" height="40" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 135px; margin-left: 1px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">SYCL-Graph Extension API</div></div></div></foreignObject><text x="220" y="139" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">SYCL-Graph Extension API</text></switch></g><rect x="0" y="195" width="440" height="40" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 215px; margin-left: 1px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">SYCL Runtime</div></div></div></foreignObject><text x="220" y="219" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">SYCL Runtime</text></switch></g><path d="M 40 315 L 40 345 L 40 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 40 353.88 L 36.5 346.88 L 40 348.63 L 43.5 346.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="0" y="275" width="440" height="40" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 295px; margin-left: 1px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Unified Runtime + Command Buffer Extension</div></div></div></foreignObject><text x="220" y="299" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Unified Runtime + Command Buffer Extension</text></switch></g><path d="M 40 395 L 40 428.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 40 433.88 L 36.5 426.88 L 40 428.63 L 43.5 426.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="0" y="355" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 375px; margin-left: 1px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">CUDA</div></div></div></foreignObject><text x="40" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">CUDA</text></switch></g><rect x="0" y="435" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 455px; margin-left: 1px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">NVIDIA GPU</div></div></div></foreignObject><text x="40" y="459" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">NVIDIA GPU</text></switch></g><path d="M 139 395 L 139 415 L 139.34 428.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 139.47 433.88 L 135.8 426.97 L 139.34 428.63 L 142.8 426.8 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="99" y="355" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 375px; margin-left: 100px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Level Zero</div></div></div></foreignObject><text x="139" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Level Zero</text></switch></g><rect x="94" y="435" width="91" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 89px; height: 1px; padding-top: 455px; margin-left: 95px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Intel CPU, GPU, FPGA, ...</div></div></div></foreignObject><text x="140" y="459" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Intel CPU, GPU,...</text></switch></g><path d="M 240 395 L 240 428.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 240 433.88 L 236.5 426.88 L 240 428.63 L 243.5 426.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="200" y="355" width="80" height="40" fill="#f8cecc" stroke="#b85450" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 375px; margin-left: 201px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">HIP</div></div></div></foreignObject><text x="240" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">HIP</text></switch></g><path d="M 370.5 395 L 370.5 415 L 370.5 414 L 370.5 427.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 370.5 432.88 L 367 425.88 L 370.5 427.63 L 374 425.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="301" y="355" width="139" height="40" fill="#fff2cc" stroke="#d6b656" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 137px; height: 1px; padding-top: 375px; margin-left: 302px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">OpenCL cl_khr_command_buffer</div></div></div></foreignObject><text x="371" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">OpenCL cl_khr_command_b...</text></switch></g><rect x="200" y="435" width="80" height="40" fill="#f8cecc" stroke="#b85450" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 455px; margin-left: 201px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">AMD GPU</div></div></div></foreignObject><text x="240" y="459" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">AMD GPU</text></switch></g><rect x="330.5" y="434" width="80" height="40" fill="#fff2cc" stroke="#d6b656" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 454px; margin-left: 332px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">CPU, GPU, FPGA, ...</div></div></div></foreignObject><text x="371" y="458" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">CPU, GPU, FPG...</text></switch></g><path d="M 220 155 L 220 188.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 220 193.88 L 216.5 186.88 L 220 188.63 L 223.5 186.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 220 235 L 220 268.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 220 273.88 L 216.5 266.88 L 220 268.63 L 223.5 266.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 239.8 315.84 L 239.97 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 239.99 353.88 L 236.46 346.9 L 239.97 348.63 L 243.46 346.86 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 139.92 314.24 L 139.14 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 139.03 353.88 L 135.68 346.81 L 139.14 348.63 L 142.68 346.96 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 370.48 315 L 370.49 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 370.49 353.88 L 366.99 346.88 L 370.49 348.63 L 373.99 346.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="10" y="10" width="430" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe flex-start; justify-content: unsafe center; width: 422px; height: 1px; padding-top: 0px; margin-left: 14px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center; max-height: 30px; overflow: hidden;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;"><div align="center"><b><font style="font-size: 20px;">SYCL-Graph Architecture</font></b></div></div></div></div></foreignObject><text x="225" y="12" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">SYCL-Graph Architecture</text></switch></g><rect x="480" y="35" width="180" height="200" fill="rgb(255, 255, 255)" stroke="rgb(0, 0, 0)" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 178px; height: 1px; padding-top: 135px; margin-left: 481px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;"><p style="line-height: 90%;" align="center"><br /></p></div></div></div></foreignObject><text x="570" y="139" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle"></text></switch></g><rect x="490" y="75" width="20" height="20" fill="#d5e8d4" stroke="#82b366" pointer-events="all"/><rect x="490" y="105" width="20" height="20" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><rect x="490" y="135" width="20" height="20" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><rect x="490" y="194" width="20" height="20" fill="#f8cecc" stroke="#b85450" pointer-events="all"/><rect x="520" y="75" width="110" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 108px; height: 1px; padding-top: 85px; margin-left: 522px;"><div style="box-sizing: border-box; font-size: 0px; text-align: left;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Application Layer</div></div></div></foreignObject><text x="522" y="89" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">Application Layer</text></switch></g><rect x="520" y="135" width="130" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 128px; height: 1px; padding-top: 145px; margin-left: 522px;"><div style="box-sizing: border-box; font-size: 0px; text-align: left;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Implemented Backend</div></div></div></foreignObject><text x="522" y="149" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">Implemented Backend</text></switch></g><rect x="520" y="105" width="110" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 108px; height: 1px; padding-top: 115px; margin-left: 522px;"><div style="box-sizing: border-box; font-size: 0px; text-align: left;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">SYCL Runtime</div></div></div></foreignObject><text x="522" y="119" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">SYCL Runtime</text></switch></g><rect x="520" y="194" width="140" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 138px; height: 1px; padding-top: 204px; margin-left: 522px;"><div style="box-sizing: border-box; font-size: 0px; text-align: left;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Future Backend Support</div></div></div></foreignObject><text x="522" y="208" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">Future Backend Support</text></switch></g><rect x="530" y="35" width="60" height="30" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 58px; height: 1px; padding-top: 50px; margin-left: 531px;"><div style="box-sizing: border-box; font-size: 0px; text-align: center;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; font-weight: bold; text-decoration: underline; white-space: normal; overflow-wrap: normal;">Legend</div></div></div></foreignObject><text x="560" y="54" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle" font-weight="bold" text-decoration="underline">Legend</text></switch></g><rect x="490" y="164" width="20" height="20" fill="#fff2cc" stroke="#d6b656" pointer-events="all"/><rect x="521" y="164" width="139" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject style="overflow: visible; text-align: left;" pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 137px; height: 1px; padding-top: 174px; margin-left: 523px;"><div style="box-sizing: border-box; font-size: 0px; text-align: left;" data-drawio-colors="color: rgb(0, 0, 0); "><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Partial Backend Support</div></div></div></foreignObject><text x="523" y="178" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">Partial Backend Support</text></switch></g></g><switch><g requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"/><a transform="translate(0,-5)" xlink:href="https://www.drawio.com/doc/faq/svg-export-text-problems" target="_blank"><text text-anchor="middle" font-size="10px" x="50%" y="100%">Text is not SVG - cannot display</text></a></switch></svg>
\ No newline at end of file
+<svg xmlns="http://www.w3.org/2000/svg" xmlns:xlink="http://www.w3.org/1999/xlink" version="1.1" width="661px" height="476px" viewBox="-0.5 -0.5 661 476" content="&lt;mxfile host=&quot;Electron&quot; modified=&quot;2024-01-02T11:08:03.405Z&quot; agent=&quot;Mozilla/5.0 (X11; Linux x86_64) AppleWebKit/537.36 (KHTML, like Gecko) draw.io/22.1.16 Chrome/120.0.6099.109 Electron/28.1.0 Safari/537.36&quot; etag=&quot;0WR0sQra9WzPLMvopOU9&quot; version=&quot;22.1.16&quot; type=&quot;device&quot;&gt;&lt;diagram name=&quot;Page-1&quot; id=&quot;rdOIFfZhXihIw1UrQana&quot;&gt;7Vtbe9o4EP01eSSf79iPhBBKm27Zpuk2fcnn2DJ4Y1uuEAT66yvZMr5IiR1iDGT3CWt0P3NmJI3EmToM12Nkx/PP0AXBmSK56zP18kxRZNXQyA+VbFKJqampYIZ8lxXKBTf+b8CEEpMufRcsSgUxhAH247LQgVEEHFyS2QjBp3IxDwblXmN7BjjBjWMHvPQf38VzNgtdyuUfgD+bZz3LEssJ7awwEyzmtgufCiJ1dKYOEYQ4/QrXQxBQ8DJc0npXz+RuB4ZAhJtUWD5erdf9px+j36ZuAaunjWbjHq1Am1nZwZLNmI0WbzIIgEsQYUmI8BzOYGQHo1x6geAycgHtRyKpvMw1hDERykT4L8B4w9RrLzEkojkOA5bLz4VNbwGXyAEvTYBxwkYzgF8ox1hH51LogCE1BjAEGG1IAQQCG/ursvZtRqLZtlyOM/lgUL8Gdg71QRwHvkO6hhGngDK8T3Mfg5vYTmB5IlZXhtLzg2AIA4iSuqqrA9PViHyBEXwEhRxTeVAN4yXwVwBhsH4RrizXYCRnVq5aafIpNxlNY0XmBXPR9gWwygF8cze87iVOishHawyiRYK1NJhOWkUcyATzvghxy+ir9p4QlzM3Uwu5tCfINSHkRPJ1GWE/BO8AZEU6NMiydQj3TBBEmx+0/rmeJe9Yc0nicl1KbVhqd7euN3TrRttuPak6QMjeFArE0I/wotDylApynmx5kbk/vbIUv648+UhHkPNkO5XdqaNz9nkb+Z5PeFEwUTIOojVpCMPQjmjOxdLzACq7zNO3Y/Pgdsxr47T2XUZDAzWPat9lcKgPby8H7W64bGB6jojRhmOCB29PGy6DZ7TZJaFNDtm/vk8uJwRbaTy9fQcIb/3DoRCWefKelsuwGrqM7Oh/JD7D4mC/BqskyvETIHiCzFZ067h8R6bwAsSTCCcQD6nvGGZOhH5cTceD9Ov8/Pwk0ddq/Yold4l+/8T9ipzFBWsdy3EFgbJxF4D/MJmeIKXVY9uMyPxu5MQorTSltHZclFY44L/EIErCQE5w/zhH9056tLx/YAfLFtnueZ7iCNnuGg+G3tJhUjPkWrbLWQC0G7rz0c7B58uT3XlX3cnhd958aLPLfUlHtLZocK+EuyCK3ynuKu9NeDceuQN6zUdSTmAvFr5TxrIMPFj7uBDHJKk7Vo5+51FMmsiCmN0HP9WGzr+p7y/oTxfoL5O9LkbKBTVV65kgW9ZEOm9WS3k+Orq196whtdJQCgzXUFuBUrXJhvgd0k5rSDv9f9rtg3Yafwpuj3ZaiXjnklJDPpKYAuSTOQH0ZmIdG2Eq4Y/qJXRjwlRvcYyOCcNv+1ojjEqOUDlhpHPLfD1hjvfW0DoyQr4XD9YkQr0jIU1Nab507sN71R/FlaNiVa13aswqvdKQ3DGrRNsxIyBoXbj+io4j8GdRkmH8WtLXdBcOUTKNLWRp8jVjv0m9h6rAg8mMc55mNWlGb5EEeOhVFmHQOq1Wabn0umiAHHrAc/ASgawHMvW0k3LHRMwNhsiSiWXSigWR8xwu20z5XBjBCFQOkUy0ICdOP5oRgZ6nviXBrJ7y3LkUkiOkFyQGO/ddF0SckWb4M9TbuX3rV543MSssXtirAuNR9nUW1UW7s1RfcUMOCugV+BHoZeOnBKNLgS5k2Ja7SMCX+Fm2vCrwsA9NGpVFSRDMkUVRBbnqr9pTJR9352CrA2oRp4+NPX9NwT3Mq0uj4pkFr1qUTm2kQbzmjcB2816oCqwsHxrZBlv9t1K2k9guh6zAHXSLLP8Qq/QUW7q2N4Jrit0X4czHBsCjLVDoSGfBgIlDssYmt08idZadeRvqMGs9yJb73aiDP0JMwjgAIV2RXJJxYTuPgL5M/G8oRGQfcrcbH377XfOu+72qQrAIdGwcoqdIs/1Yw3YL+AYl0ANPdsWut6ERq3YraQj0UY2C7KKPsf93eP/r8dMHbzr+uPrkTX7+6Av+ZNL2otzN1R+3KBv8xd++FmUhsPx969QmLLSD4gIg3SzjGCL8fv2PXKsV4TuDNtSif11/9y9H6m3/o6GuvoST39+UnsxB3cV/T3a4O0vHWQnE7Rj9Y2DW34pJYm2+9eENSeb/EU1javk/bdXRHw==&lt;/diagram&gt;&lt;/mxfile&gt;" style="background-color: rgb(255, 255, 255);"><defs/><g><path d="M 220 75 L 220 108.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 220 113.88 L 216.5 106.88 L 220 108.63 L 223.5 106.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="0" y="34" width="440" height="41" fill="#d5e8d4" stroke="#82b366" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 55px; margin-left: 1px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Application</div></div></div></foreignObject><text x="220" y="58" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Application</text></switch></g><rect x="0" y="115" width="440" height="40" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 135px; margin-left: 1px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">SYCL-Graph Extension API</div></div></div></foreignObject><text x="220" y="139" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">SYCL-Graph Extension API</text></switch></g><rect x="0" y="195" width="440" height="40" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 215px; margin-left: 1px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">SYCL Runtime</div></div></div></foreignObject><text x="220" y="219" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">SYCL Runtime</text></switch></g><path d="M 40 315 L 40 345 L 40 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 40 353.88 L 36.5 346.88 L 40 348.63 L 43.5 346.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="0" y="275" width="440" height="40" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 438px; height: 1px; padding-top: 295px; margin-left: 1px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Unified Runtime + Command Buffer Extension</div></div></div></foreignObject><text x="220" y="299" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Unified Runtime + Command Buffer Extension</text></switch></g><path d="M 40 395 L 40 428.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 40 433.88 L 36.5 426.88 L 40 428.63 L 43.5 426.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="0" y="355" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 375px; margin-left: 1px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">CUDA</div></div></div></foreignObject><text x="40" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">CUDA</text></switch></g><rect x="0" y="435" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 455px; margin-left: 1px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">NVIDIA GPU</div></div></div></foreignObject><text x="40" y="459" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">NVIDIA GPU</text></switch></g><path d="M 139 395 L 139 415 L 139.34 428.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 139.47 433.88 L 135.8 426.97 L 139.34 428.63 L 142.8 426.8 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="99" y="355" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 375px; margin-left: 100px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Level Zero</div></div></div></foreignObject><text x="139" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Level Zero</text></switch></g><rect x="94" y="435" width="91" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 89px; height: 1px; padding-top: 455px; margin-left: 95px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Intel CPU, GPU, FPGA, ...</div></div></div></foreignObject><text x="140" y="459" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">Intel CPU, GPU,...</text></switch></g><path d="M 240 395 L 240 428.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 240 433.88 L 236.5 426.88 L 240 428.63 L 243.5 426.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="200" y="355" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 375px; margin-left: 201px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">HIP</div></div></div></foreignObject><text x="240" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">HIP</text></switch></g><path d="M 370.5 395 L 370.5 415 L 370.5 414 L 370.5 427.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 370.5 432.88 L 367 425.88 L 370.5 427.63 L 374 425.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="301" y="355" width="139" height="40" fill="#fff2cc" stroke="#d6b656" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 137px; height: 1px; padding-top: 375px; margin-left: 302px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">OpenCL cl_khr_command_buffer</div></div></div></foreignObject><text x="371" y="379" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">OpenCL cl_khr_command_b...</text></switch></g><rect x="200" y="435" width="80" height="40" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 455px; margin-left: 201px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">AMD GPU</div></div></div></foreignObject><text x="240" y="459" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">AMD GPU</text></switch></g><rect x="330.5" y="434" width="80" height="40" fill="#fff2cc" stroke="#d6b656" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 78px; height: 1px; padding-top: 454px; margin-left: 332px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">CPU, GPU, FPGA, ...</div></div></div></foreignObject><text x="371" y="458" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">CPU, GPU, FPG...</text></switch></g><path d="M 220 155 L 220 188.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 220 193.88 L 216.5 186.88 L 220 188.63 L 223.5 186.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 220 235 L 220 268.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 220 273.88 L 216.5 266.88 L 220 268.63 L 223.5 266.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 239.8 315.84 L 239.97 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 239.99 353.88 L 236.46 346.9 L 239.97 348.63 L 243.46 346.86 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 139.92 314.24 L 139.14 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 139.03 353.88 L 135.68 346.81 L 139.14 348.63 L 142.68 346.96 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><path d="M 370.48 315 L 370.49 348.63" fill="none" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="stroke"/><path d="M 370.49 353.88 L 366.99 346.88 L 370.49 348.63 L 373.99 346.88 Z" fill="rgb(0, 0, 0)" stroke="rgb(0, 0, 0)" stroke-miterlimit="10" pointer-events="all"/><rect x="10" y="10" width="430" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe flex-start; justify-content: unsafe center; width: 422px; height: 1px; padding-top: 0px; margin-left: 14px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center; max-height: 30px; overflow: hidden;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;"><div align="center"><b><font style="font-size: 20px;">SYCL-Graph Architecture</font></b></div></div></div></div></foreignObject><text x="225" y="12" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle">SYCL-Graph Architecture</text></switch></g><rect x="480" y="35" width="180" height="160" fill="rgb(255, 255, 255)" stroke="rgb(0, 0, 0)" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 178px; height: 1px; padding-top: 115px; margin-left: 481px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;"><p style="line-height: 90%;" align="center"><br /></p></div></div></div></foreignObject><text x="570" y="119" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle"></text></switch></g><rect x="490" y="75" width="20" height="20" fill="#d5e8d4" stroke="#82b366" pointer-events="all"/><rect x="490" y="105" width="20" height="20" fill="#e1d5e7" stroke="#9673a6" pointer-events="all"/><rect x="490" y="135" width="20" height="20" fill="#dae8fc" stroke="#6c8ebf" pointer-events="all"/><rect x="520" y="75" width="110" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 108px; height: 1px; padding-top: 85px; margin-left: 522px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: left;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Application Layer</div></div></div></foreignObject><text x="522" y="89" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">Application Layer</text></switch></g><rect x="520" y="135" width="130" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 128px; height: 1px; padding-top: 145px; margin-left: 522px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: left;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Implemented Backend</div></div></div></foreignObject><text x="522" y="149" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">Implemented Backend</text></switch></g><rect x="520" y="105" width="110" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 108px; height: 1px; padding-top: 115px; margin-left: 522px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: left;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">SYCL Runtime</div></div></div></foreignObject><text x="522" y="119" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">SYCL Runtime</text></switch></g><rect x="530" y="35" width="60" height="30" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe center; width: 58px; height: 1px; padding-top: 50px; margin-left: 531px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: center;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; font-weight: bold; text-decoration: underline; white-space: normal; overflow-wrap: normal;">Legend</div></div></div></foreignObject><text x="560" y="54" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px" text-anchor="middle" font-weight="bold" text-decoration="underline">Legend</text></switch></g><rect x="490" y="164" width="20" height="20" fill="#fff2cc" stroke="#d6b656" pointer-events="all"/><rect x="521" y="164" width="139" height="20" fill="none" stroke="none" pointer-events="all"/><g transform="translate(-0.5 -0.5)"><switch><foreignObject pointer-events="none" width="100%" height="100%" requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility" style="overflow: visible; text-align: left;"><div xmlns="http://www.w3.org/1999/xhtml" style="display: flex; align-items: unsafe center; justify-content: unsafe flex-start; width: 137px; height: 1px; padding-top: 174px; margin-left: 523px;"><div data-drawio-colors="color: rgb(0, 0, 0); " style="box-sizing: border-box; font-size: 0px; text-align: left;"><div style="display: inline-block; font-size: 12px; font-family: Helvetica; color: rgb(0, 0, 0); line-height: 1.2; pointer-events: all; white-space: normal; overflow-wrap: normal;">Partial Backend Support</div></div></div></foreignObject><text x="523" y="178" fill="rgb(0, 0, 0)" font-family="Helvetica" font-size="12px">Partial Backend Support</text></switch></g></g><switch><g requiredFeatures="http://www.w3.org/TR/SVG11/feature#Extensibility"/><a transform="translate(0,-5)" xlink:href="https://www.drawio.com/doc/faq/svg-export-text-problems" target="_blank"><text text-anchor="middle" font-size="10px" x="50%" y="100%">Text is not SVG - cannot display</text></a></switch></svg>
\ No newline at end of file

From 156b4b7d63843074348e3d7d3e9720ca72c148e1 Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Mon, 15 Jan 2024 17:04:01 +0000
Subject: [PATCH 07/12] Adds debug print

---
 sycl/test-e2e/Graph/Inputs/buffer_fill.cpp | 4 ++++
 1 file changed, 4 insertions(+)

diff --git a/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp b/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp
index 351194dadda0f..5303df20436ff 100644
--- a/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp
+++ b/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp
@@ -78,6 +78,10 @@ int main() {
   host_accessor HostDataI8(BufferI8);
   for (int i = 0; i < N; i++) {
     assert(HostData[i] == Pattern);
+    if (HostDataI64[i] != PatternI64) {
+      std::cout << "HostDataI64[" << i << "] = " << HostDataI64[i]
+                << " == PatternI64 = " << PatternI64 << std::endl;
+    }
     assert(HostDataI64[i] == PatternI64);
     assert(HostDataI32[i] == PatternI32);
     assert(HostDataI16[i] == PatternI16);

From 24b67d818a54cc5dc5791068c8d3883c7ac7edfe Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Tue, 16 Jan 2024 10:20:30 +0000
Subject: [PATCH 08/12] Removes debug print

---
 sycl/test-e2e/Graph/Inputs/buffer_fill.cpp | 4 ----
 1 file changed, 4 deletions(-)

diff --git a/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp b/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp
index 5303df20436ff..351194dadda0f 100644
--- a/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp
+++ b/sycl/test-e2e/Graph/Inputs/buffer_fill.cpp
@@ -78,10 +78,6 @@ int main() {
   host_accessor HostDataI8(BufferI8);
   for (int i = 0; i < N; i++) {
     assert(HostData[i] == Pattern);
-    if (HostDataI64[i] != PatternI64) {
-      std::cout << "HostDataI64[" << i << "] = " << HostDataI64[i]
-                << " == PatternI64 = " << PatternI64 << std::endl;
-    }
     assert(HostDataI64[i] == PatternI64);
     assert(HostDataI32[i] == PatternI32);
     assert(HostDataI16[i] == PatternI16);

From 05d24936d1b32ef28ebf012d088beb75d2a2bc08 Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Tue, 16 Jan 2024 12:13:32 +0000
Subject: [PATCH 09/12] Update UR CMakeFile

---
 sycl/plugins/unified_runtime/CMakeLists.txt | 2 +-
 1 file changed, 1 insertion(+), 1 deletion(-)

diff --git a/sycl/plugins/unified_runtime/CMakeLists.txt b/sycl/plugins/unified_runtime/CMakeLists.txt
index c651a04e2a469..7ffacb41a766f 100644
--- a/sycl/plugins/unified_runtime/CMakeLists.txt
+++ b/sycl/plugins/unified_runtime/CMakeLists.txt
@@ -57,7 +57,7 @@ if(SYCL_PI_UR_USE_FETCH_CONTENT)
   include(FetchContent)
   
   set(UNIFIED_RUNTIME_REPO "https://github.com/bensuo/unified-runtime.git")
-  set(UNIFIED_RUNTIME_TAG cmdbuf-hip-support)
+  set(UNIFIED_RUNTIME_TAG cmdbuf-support-hip)
 
   if(SYCL_PI_UR_OVERRIDE_FETCH_CONTENT_REPO)
     set(UNIFIED_RUNTIME_REPO "${SYCL_PI_UR_OVERRIDE_FETCH_CONTENT_REPO}")

From 0a2c4852feed37b7429989cd1c7e889b392fa09d Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Fri, 19 Jan 2024 12:26:16 +0000
Subject: [PATCH 10/12] Fixes merge issue

---
 sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp     | 2 +-
 sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp | 2 +-
 2 files changed, 2 insertions(+), 2 deletions(-)

diff --git a/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp b/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
index 505b5e78c49eb..f41d29264d6f7 100644
--- a/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
+++ b/sycl/test-e2e/Graph/Explicit/work_group_size_prop.cpp
@@ -10,7 +10,7 @@
 // should. So the Sycl graph support cannot correctly catch the error and throw
 // the approriate exception for negative test. An issue has been reported
 // https://github.com/bashbaug/SimpleOpenCLSamples/issues/95
-// XFAIL: cuda
+// XFAIL: cuda, hip
 // UNSUPPORTED: opencl
 // Note: failing negative test with HIP in the original test
 
diff --git a/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp b/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
index 2970b7efa3260..ae5929d9541e7 100644
--- a/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
+++ b/sycl/test-e2e/Graph/RecordReplay/work_group_size_prop.cpp
@@ -10,7 +10,7 @@
 // should. So the Sycl graph support cannot correctly catch the error and throw
 // the approriate exception for negative test. An issue has been reported
 // https://github.com/bashbaug/SimpleOpenCLSamples/issues/95
-// XFAIL: cuda
+// XFAIL: cuda, hip
 // UNSUPPORTED: opencl
 // Note: failing negative test with HIP in the original test
 

From 6dea501db095579769b0092e4244f8689cb8d304 Mon Sep 17 00:00:00 2001
From: Ewan Crawford <ewan@codeplay.com>
Date: Mon, 19 Feb 2024 12:42:19 +0000
Subject: [PATCH 11/12] Check for shared allocation support

---
 sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp | 4 ++++
 1 file changed, 4 insertions(+)

diff --git a/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp b/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp
index 04da3eaa5414b..46e3c2b850007 100644
--- a/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp
+++ b/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp
@@ -18,6 +18,10 @@ int main() {
     return 0;
   }
 
+  if (!Queue.get_device().has(sycl::aspect::usm_shared_allocations)) {
+    return 0;
+  }
+
   exp_ext::command_graph Graph{Queue.get_context(), Queue.get_device()};
 
   const size_t N = 10;

From de18a505acb24bd2ec9ecc5a3f0f22838b35870f Mon Sep 17 00:00:00 2001
From: Maxime France-Pillois <maxime.francepillois@codeplay.com>
Date: Thu, 29 Feb 2024 12:22:29 +0000
Subject: [PATCH 12/12] Add REQUIRES to check usm_shared_allocations aspect

---
 sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp | 5 +----
 1 file changed, 1 insertion(+), 4 deletions(-)

diff --git a/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp b/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp
index 46e3c2b850007..da3ccd2329ec9 100644
--- a/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp
+++ b/sycl/test-e2e/Graph/Explicit/enqueue_ordering.cpp
@@ -1,3 +1,4 @@
+// REQUIRES: aspect-usm_shared_allocations
 // RUN: %{build} -o %t.out
 // RUN: %{run} %t.out
 // Extra run to check for leaks in Level Zero using UR_L0_LEAKS_DEBUG
@@ -18,10 +19,6 @@ int main() {
     return 0;
   }
 
-  if (!Queue.get_device().has(sycl::aspect::usm_shared_allocations)) {
-    return 0;
-  }
-
   exp_ext::command_graph Graph{Queue.get_context(), Queue.get_device()};
 
   const size_t N = 10;