[SYCL] Tests for Level Zero linker flags (#713)

The Level Zero backend never supported linker options, but it
previously ignored them silently. The backend now raises an error if a
request to link kernel code specifies any linker options.

Add two new tests to verify this behavior. One test exercises the
deprecated program API. The other exercises the new kernel_bundle
API. The test using the kernel_bundle API is currently disabled (see
the comments in the test for why).

Also mark two existing tests as UNSUPPORTED on Level Zero. These tests
were incorrectly passing OpenCL linker flags to the Level Zero backend.
This used to work when the backend silently ignored the flags, but it
now raises an error.

Author: gmlueck

SYCL/DeprecatedFeatures/get-options.cpp

@@ -3,7 +3,10 @@
 // RUN: %CPU_RUN_PLACEHOLDER %t.out
 // RUN: %GPU_RUN_PLACEHOLDER %t.out
 // RUN: %ACC_RUN_PLACEHOLDER %t.out
-// XFAIL: cuda || hip
+//
+// This test passes OpenCL specific compiler and linker swiches to the backend,
+// so it is unsupported on any other backend.
+// UNSUPPORTED: cuda || hip || level_zero
 
 #include <CL/sycl.hpp>
 

SYCL/DeprecatedFeatures/level-zero-link-flags.cpp

@@ -0,0 +1,44 @@
+// RUN: %clangxx -fsycl -D__SYCL_INTERNAL_API %s -o %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+// REQUIRES: level_zero
+//
+//==--- level-zero-link-flags.cpp - Error handling for link flags --==//
+//
+// The Level Zero backend does not accept any online linker options.
+// This test validates that an error is raised if you attempt to pass any.
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===--------------------------------------------------------------===//
+
+#include <CL/sycl.hpp>
+
+class MyKernel;
+
+void test() {
+  sycl::queue Queue;
+  sycl::context Context = Queue.get_context();
+
+  sycl::program Program(Context);
+  Program.compile_with_kernel_type<MyKernel>();
+
+  try {
+    Program.link("-foo");
+    assert(false && "Expected error linking program");
+  } catch (const sycl::exception &e) {
+    assert((e.code() == sycl::errc::build) && "Wrong error code");
+  } catch (...) {
+    assert(false && "Expected sycl::exception");
+  }
+
+  Queue.submit([&](sycl::handler &CGH) { CGH.single_task<MyKernel>([=] {}); })
+      .wait();
+}
+
+int main() {
+  test();
+
+  return 0;
+}

SYCL/DeprecatedFeatures/program_link.cpp

@@ -9,6 +9,10 @@
 //
 // Hits an assertion on AMD with multiple GPUs available, fails trace on Nvidia.
 // XFAIL: hip_amd || hip_nvidia
+//
+// Unsupported on Level Zero because the test passes OpenCL specific compiler
+// and linker switches.
+// UNSUPPORTED: level_zero
 
 #include <CL/sycl.hpp>
 #include <iostream>

SYCL/KernelAndProgram/level-zero-link-flags.cpp

@@ -0,0 +1,55 @@
+// RUN: %clangxx -fsycl -Xsycl-target-linker=spir64 -foo %s -o %t.out
+// RUN: %GPU_RUN_PLACEHOLDER %t.out
+// REQUIRES: level_zero
+//
+// This test is disabled because the runtime does not currently pass linker
+// flags from "-Xsycl-target-linker=spir64 <opts>" when the program calls
+// "sycl::link()".  This seems like a bug.  Note that the runtime does pass
+// those "<opts>" to the online linker in other cases when it links device
+// code.
+//
+// XFAIL: level_zero
+//
+//==--- level-zero-link-flags.cpp - Error handling for link flags --==//
+//
+// The Level Zero backend does not accept any online linker options.
+// This test validates that an error is raised if you attempt to pass any.
+//
+// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// See https://llvm.org/LICENSE.txt for license information.
+// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
+//
+//===--------------------------------------------------------------===//
+
+#include <sycl/sycl.hpp>
+
+class MyKernel;
+
+void test() {
+  sycl::queue Queue;
+  sycl::context Context = Queue.get_context();
+
+  auto BundleInput =
+      sycl::get_kernel_bundle<MyKernel, sycl::bundle_state::input>(Context);
+  auto BundleObject = sycl::compile(BundleInput);
+
+  try {
+    sycl::link(BundleObject);
+    assert(false && "Expected error linking kernel bundle");
+  } catch (const sycl::exception &e) {
+    std::string Msg(e.what());
+    assert((e.code() == sycl::errc::build) && "Wrong error code");
+    assert(Msg.find("-foo") != std::string::npos);
+  } catch (...) {
+    assert(false && "Expected sycl::exception");
+  }
+
+  Queue.submit([&](sycl::handler &CGH) { CGH.single_task<MyKernel>([=] {}); })
+      .wait();
+}
+
+int main() {
+  test();
+
+  return 0;
+}