Merge remote-tracking branch 'origin/sycl' into abi/get_backend_info

Author: againull

.github/workflows/sycl-linux-build.yml

@@ -204,7 +204,8 @@ jobs:
           --ci-defaults --use-zstd ${{ inputs.build_configure_extra_args }} \
           -DCMAKE_C_COMPILER_LAUNCHER=ccache \
           -DCMAKE_CXX_COMPILER_LAUNCHER=ccache \
-          -DLLVM_INSTALL_UTILS=ON
+          -DLLVM_INSTALL_UTILS=ON \
+          -DSYCL_UR_FORCE_FETCH_LEVEL_ZERO=ON
     - name: Compile
       id: build
       # Emulate default value for manual dispatch as we've run out of available arguments.

clang/include/clang/Basic/DiagnosticSemaKinds.td

@@ -13425,8 +13425,9 @@ def err_sycl_add_ir_attribute_invalid_filter : Error<
   "initializer list in the first argument of %0 must contain only string "
   "literals">;
 def warn_sycl_old_and_new_kernel_attributes : Warning<
-  "kernel has both attribute %0 and kernel properties; conflicting properties "
-  "are ignored">, InGroup<IgnoredAttributes>;
+  "kernel has both attribute %0 and kernel properties; if the kernel "
+  "properties contains the property \"%1\" it will be ignored">,
+  InGroup<IgnoredAttributes>;
 def err_attribute_argument_is_not_valid : Error<
   "%0 attribute requires integer constant value 0 or 1">;
 

clang/lib/Sema/SemaSYCLDeclAttr.cpp

@@ -3256,14 +3256,19 @@ void SemaSYCL::checkSYCLAddIRAttributesFunctionAttrConflicts(Decl *D) {
   }
 
   // If there are potentially conflicting attributes, we issue a warning.
-  for (const auto *Attr : std::vector<AttributeCommonInfo *>{
-           D->getAttr<SYCLReqdWorkGroupSizeAttr>(),
-           D->getAttr<IntelReqdSubGroupSizeAttr>(),
-           D->getAttr<SYCLWorkGroupSizeHintAttr>(),
-           D->getAttr<SYCLDeviceHasAttr>()})
+  for (const auto [Attr, PotentialConflictProp] :
+       std::vector<std::pair<AttributeCommonInfo *, StringRef>>{
+           {D->getAttr<SYCLReqdWorkGroupSizeAttr>(),
+            "sycl::ext::oneapi::experimental::work_group_size"},
+           {D->getAttr<IntelReqdSubGroupSizeAttr>(),
+            "sycl::ext::oneapi::experimental::sub_group_size"},
+           {D->getAttr<SYCLWorkGroupSizeHintAttr>(),
+            "sycl::ext::oneapi::experimental::work_group_size_hint"},
+           {D->getAttr<SYCLDeviceHasAttr>(),
+            "sycl::ext::oneapi::experimental::device_has"}})
     if (Attr)
       Diag(Attr->getLoc(), diag::warn_sycl_old_and_new_kernel_attributes)
-          << Attr;
+          << Attr << PotentialConflictProp;
 }
 
 void SemaSYCL::handleSYCLRegisteredKernels(Decl *D, const ParsedAttr &A) {

clang/test/Driver/clang-linker-wrapper.cpp

@@ -63,13 +63,20 @@
 // CHK-CMDS-AOT-GEN: spirv-to-ir-wrapper{{.*}} -o [[FIRSTLLVMLINKIN:.*]].bc --llvm-spirv-opts --spirv-preserve-auxdata --spirv-target-env=SPV-IR --spirv-builtin-format=global
 // CHK-CMDS-AOT-GEN-NEXT: llvm-link{{.*}} --suppress-warnings [[FIRSTLLVMLINKIN]].bc -o [[FIRSTLLVMLINKOUT:.*]].bc
 // CHK-CMDS-AOT-GEN-NEXT: llvm-link{{.*}} -only-needed --suppress-warnings [[FIRSTLLVMLINKOUT]].bc {{.*}}.bc -o [[SECONDLLVMLINKOUT:.*]].bc
-// CHK-CMDS-AOT-GEN-NEXT: sycl-post-link{{.*}} SYCL_POST_LINK_OPTIONS -o [[SYCLPOSTLINKOUT:.*]].table [[SECONDLLVMLINKOUT]].bc
+// Check that target specified by -fsycl-targets is passed to sycl-post-link for filtering.
+// CHK-CMDS-AOT-GEN-NEXT: sycl-post-link{{.*}} SYCL_POST_LINK_OPTIONS -o intel_gpu_pvc,[[SYCLPOSTLINKOUT:.*]].table [[SECONDLLVMLINKOUT]].bc
 // CHK-CMDS-AOT-GEN-NEXT: llvm-spirv{{.*}} LLVM_SPIRV_OPTIONS -o {{.*}}
 // CHK-CMDS-AOT-GEN-NEXT: ocloc{{.*}} -output_no_suffix -spirv_input -device pvc{{.*}} -output {{.*}} -file {{.*}}
 // CHK-CMDS-AOT-GEN-NEXT: offload-wrapper: output: [[WRAPPEROUT:.*]].bc, input: {{.*}}, compile-opts: , link-opts:
 // CHK-CMDS-AOT-GEN-NEXT: clang{{.*}} -c -o [[LLCOUT:.*]].o [[WRAPPEROUT]].bc
 // CHK-CMDS-AOT-GEN-NEXT: "{{.*}}/ld" -- HOST_LINKER_FLAGS -dynamic-linker HOST_DYN_LIB -o a.out [[LLCOUT]].o HOST_LIB_PATH HOST_STAT_LIB {{.*}}.o
 
+// Check that when --gpu-tool-arg is specified in clang-linker-wrapper 
+// (happen when AOT device is specified via -Xsycl-target-backend '-device pvc' in clang), 
+// the target is not passed to sycl-post-link for filtering.
+// RUN: clang-linker-wrapper -sycl-embed-ir -sycl-device-libraries=%t1.devicelib.o -sycl-post-link-options="SYCL_POST_LINK_OPTIONS" -llvm-spirv-options="LLVM_SPIRV_OPTIONS" "--host-triple=x86_64-unknown-linux-gnu" "--gpu-tool-arg=-device pvc" "--linker-path=/usr/bin/ld" "--" HOST_LINKER_FLAGS "-dynamic-linker" HOST_DYN_LIB "-o" "a.out" HOST_LIB_PATH HOST_STAT_LIB %t1.o --dry-run 2>&1 | FileCheck -check-prefix=CHK-NO-CMDS-AOT-GEN %s
+// CHK-NO-CMDS-AOT-GEN-NOT: sycl-post-link{{.*}} -o intel_gpu_pv,{{.*}}
+
 /// Check for list of commands for standalone clang-linker-wrapper run for sycl (AOT for Intel CPU)
 // -------
 // Generate .o file as linker wrapper input.