[SYCLomatic] Use the new interface of sycl::queue::ext_oneapi_get_last_event() and keep the code can be built by both new and old oneAPI compiler (#2716)

Signed-off-by: Jiang, Zhiwei <[email protected]>

Author: zhiweij1

clang/lib/DPCT/AnalysisInfo.cpp

@@ -6667,19 +6667,25 @@ KernelPrinter &KernelCallExpr::SubmitStmtsList::print(KernelPrinter &Printer) {
       Printer.line("cgh.depends_on(dpct::get_current_device().get_in_order_"
                    "queues_last_events());");
     } else {
-      Printer.line("#ifdef __INTEL_LLVM_COMPILER");
-      Printer.newLine();
-      Printer.line("cgh.depends_on(dpct::get_default_queue().ext_oneapi_get_"
-                   "last_event());");
-      Printer.newLine();
-      Printer.line("#else");
-      Printer.newLine();
-      Printer.line("auto e_opt = dpct::get_default_queue().ext_oneapi_get_last_"
-                   "event();");
-      Printer.newLine();
-      Printer.line("if (e_opt) cgh.depends_on(*e_opt);");
-      Printer.newLine();
-      Printer.line("#endif");
+      Printer.line("auto last_event = "
+                   "dpct::get_default_queue().ext_oneapi_get_last_event();");
+      Printer.line("[&](auto &&_e) {");
+      {
+        Printer.indent();
+        Printer.line("if constexpr "
+                     "(std::is_same_v<std::remove_reference_t<decltype(last_"
+                     "event)>, sycl::event>)");
+        {
+          Printer.indent();
+          Printer.line("cgh.depends_on(_e);");
+        }
+        Printer.line("else if (_e.has_value())");
+        {
+          Printer.indent();
+          Printer.line("cgh.depends_on(_e.value());");
+        }
+      }
+      Printer.line("}(last_event);");
     }
     Printer.newLine();
   }

clang/runtime/dpct-rt/include/dpct/device.hpp

@@ -592,14 +592,15 @@ class device_ext : public sycl::device {
     lock.unlock();
     for (const auto &q : current_queues) {
       if (q->is_in_order()) {
-#ifdef __INTEL_LLVM_COMPILER
-        last_events.push_back(q->ext_oneapi_get_last_event());
-#else
         auto last_event = q->ext_oneapi_get_last_event();
-        if (last_event) {
-          last_events.push_back(*last_event);
-        }
-#endif
+        [&](auto &&_e) {
+          if constexpr (std::is_same_v<
+                            std::remove_reference_t<decltype(last_event)>,
+                            sycl::event>)
+            last_events.push_back(_e);
+          else if (_e.has_value())
+            last_events.push_back(_e.value());
+        }(last_event);
       }
     }
     // Guard the destruct of current_queues to make sure the ref count is safe.

clang/test/dpct/kernel_implicit_sync.cu

@@ -27,12 +27,13 @@ int main() {
 
 // CHECK:  s1->submit(
 // CHECK:      [&](sycl::handler &cgh) {
-// CHECK:#ifdef __INTEL_LLVM_COMPILER
-// CHECK:        cgh.depends_on(dpct::get_default_queue().ext_oneapi_get_last_event());
-// CHECK:#else
-// CHECK:        auto e_opt = dpct::get_default_queue().ext_oneapi_get_last_event();
-// CHECK:        if (e_opt) cgh.depends_on(*e_opt);
-// CHECK:#endif
+// CHECK:        auto last_event = dpct::get_default_queue().ext_oneapi_get_last_event();
+// CHECK:        [&](auto &&_e) {
+// CHECK:          if constexpr (std::is_same_v<std::remove_reference_t<decltype(last_event)>, sycl::event>)
+// CHECK:            cgh.depends_on(_e);
+// CHECK:          else if (_e.has_value())
+// CHECK:            cgh.depends_on(_e.value());
+// CHECK:        }(last_event);
 // CHECK:        cgh.parallel_for(
 // CHECK:          sycl::nd_range<3>(sycl::range<3>(1, 1, 1), sycl::range<3>(1, 1, 1)),
 // CHECK:          [=](sycl::nd_item<3> item_ct1) {