From 914561a7eefa3d7270a95c5b80cbe9c280e17c74 Mon Sep 17 00:00:00 2001
From: Sergey Semenov <sergey.semenov@intel.com>
Date: Thu, 1 Aug 2024 01:43:19 -0700
Subject: [PATCH] [SYCL] Fix dependency compilation after PI removal (#14843)

---
 .../program_manager/program_manager.cpp       | 11 +++-----
 .../program_manager/DynamicLinking.cpp        | 26 +++++++++++++++++++
 2 files changed, 29 insertions(+), 8 deletions(-)

diff --git a/sycl/source/detail/program_manager/program_manager.cpp b/sycl/source/detail/program_manager/program_manager.cpp
index 92c5e9ed8f132..7de12004dd572 100644
--- a/sycl/source/detail/program_manager/program_manager.cpp
+++ b/sycl/source/detail/program_manager/program_manager.cpp
@@ -1507,14 +1507,9 @@ ProgramManager::ProgramPtr ProgramManager::build(
   LinkPrograms.push_back(Program.get());
 
   for (ur_program_handle_t Prg : ExtraProgramsToLink) {
-    auto Result =
-        Plugin->call_nocheck(urProgramCompileExp, Prg, /* num devices =*/1,
-                             &Device, CompileOptions.c_str());
-    if (Result == UR_RESULT_ERROR_UNSUPPORTED_FEATURE) {
-      Plugin->call(urProgramCompile, Context->getHandleRef(), Prg,
-                   CompileOptions.c_str());
-    }
-    Plugin->checkUrResult(Result);
+    auto Res = doCompile(Plugin, Prg, /*num devices =*/1, &Device,
+                         Context->getHandleRef(), CompileOptions.c_str());
+    Plugin->checkUrResult(Res);
 
     LinkPrograms.push_back(Prg);
   }
diff --git a/sycl/unittests/program_manager/DynamicLinking.cpp b/sycl/unittests/program_manager/DynamicLinking.cpp
index c8c9b101baf9e..7c86296afab5b 100644
--- a/sycl/unittests/program_manager/DynamicLinking.cpp
+++ b/sycl/unittests/program_manager/DynamicLinking.cpp
@@ -197,4 +197,30 @@ TEST(DynamicLinking, AheadOfTime) {
   }
 }
 
+static ur_result_t redefined_urProgramCompileExp(void *pParams) {
+  return UR_RESULT_ERROR_UNSUPPORTED_FEATURE;
+}
+
+TEST(DynamicLinking, UnsupportedCompileExp) {
+  sycl::unittest::UrMock<> Mock;
+  setupRuntimeLinkingMock();
+  mock::getCallbacks().set_replace_callback("urProgramCompileExp",
+                                            redefined_urProgramCompileExp);
+
+  sycl::platform Plt = sycl::platform();
+  sycl::queue Q(Plt.get_devices()[0]);
+
+  CapturedLinkingData.clear();
+
+  Q.single_task<DynamicLinkingTest::BasicCaseKernel>([=]() {});
+  ASSERT_EQ(CapturedLinkingData.NumOfUrProgramCreateCalls, 3u);
+  // Both programs should be linked together.
+  ASSERT_EQ(CapturedLinkingData.NumOfUrProgramLinkCalls, 1u);
+  ASSERT_TRUE(CapturedLinkingData.LinkedProgramsContains(
+      {BASIC_CASE_PRG, BASIC_CASE_PRG_DEP, BASIC_CASE_PRG_DEP_DEP}));
+  // And the linked program should be used to create a kernel.
+  ASSERT_EQ(CapturedLinkingData.ProgramUsedToCreateKernel,
+            BASIC_CASE_PRG * BASIC_CASE_PRG_DEP * BASIC_CASE_PRG_DEP_DEP);
+}
+
 } // anonymous namespace