[SYCL][sycl-post-link] Makes device_global property names consistent (#5903)

This commit changes the IR attributes expected for device_global
properties more consistent with other sycl-related IR attributes, i.e.
the naming being prefixed by "sycl-" and replacing "_" with "-".

Signed-off-by: Steffen Larsen <[email protected]>

Author: steffenlarsen

llvm/test/tools/sycl-post-link/device-globals/test_global_variable.ll

@@ -6,7 +6,7 @@
 ; This test is intended to check that DeviceGlobalPass adds all the required
 ; metadata nodes to every device global variable as well as the required
 ; properties in the 'SYCL/device globals' property set and handles the
-; 'device_image_scope' attribute written in any allowed form.
+; 'sycl-device-image-scope' attribute written in any allowed form.
 
 source_filename = "test_global_variable.cpp"
 target datalayout = "e-i64:64-v16:16-v24:32-v32:32-v48:64-v96:128-v192:256-v256:256-v512:512-v1024:1024-n8:16:32:64"
@@ -46,15 +46,15 @@ declare spir_func align 4 dereferenceable(4) i32 addrspace(4)* @_ZNK2cl4sycl3ext
 ; Function Attrs: convergent mustprogress noinline norecurse nounwind optnone
 declare spir_func align 1 dereferenceable(1) i8 addrspace(4)* @_ZNK2cl4sycl3ext6oneapi13device_globalIbJNS2_8PropertyIXadsoKcL_ZL5Name1EEEXadsoS5_L_ZL6Value1EEEEENS4_IXadsoS5_L_ZL5Name2EEEXadsoS5_L_ZL6Value2EEEEENS4_IXadsoS5_L_ZL5Name3EEEXadsoS5_L_ZL6Value3EEEEENS4_IXadsoS5_L_ZL5Name4EEEXadsoS5_L_ZL6Value4EEEEEEE3getEv(%"class.cl::sycl::ext::oneapi::device_global.1" addrspace(4)* align 1 dereferenceable_or_null(1) %this) #4 align 2
 
-attributes #0 = { "sycl-unique-id"="6da74a122db9f35d____ZL7dg_int1" "device_image_scope"="false" "host_access"="1" "implement_in_csr"="true" "init_mode"="0" "sycl-device-global-size"="4" }
-attributes #1 = { "sycl-unique-id"="7da74a1187b9f35d____ZL7dg_int2" "implement_in_csr"="false" "init_mode"="1" "sycl-device-global-size"="4" }
-attributes #2 = { "sycl-unique-id"="9d329ad59055e972____ZL8dg_bool3" "device_image_scope"="true" "host_access"="0" "implement_in_csr" "init_mode"="0" "sycl-device-global-size"="1" }
-attributes #3 = { "sycl-unique-id"="dda2bad52c45c432____ZL8dg_bool4" "device_image_scope" "host_access"="2" "sycl-device-global-size"="1" }
+attributes #0 = { "sycl-unique-id"="6da74a122db9f35d____ZL7dg_int1" "sycl-device-image-scope"="false" "sycl-host-access"="1" "sycl-implement-in-csr"="true" "sycl-init-mode"="0" "sycl-device-global-size"="4" }
+attributes #1 = { "sycl-unique-id"="7da74a1187b9f35d____ZL7dg_int2" "sycl-implement-in-csr"="false" "sycl-init-mode"="1" "sycl-device-global-size"="4" }
+attributes #2 = { "sycl-unique-id"="9d329ad59055e972____ZL8dg_bool3" "sycl-device-image-scope"="true" "sycl-host-access"="0" "sycl-implement-in-csr" "sycl-init-mode"="0" "sycl-device-global-size"="1" }
+attributes #3 = { "sycl-unique-id"="dda2bad52c45c432____ZL8dg_bool4" "sycl-device-image-scope" "sycl-host-access"="2" "sycl-device-global-size"="1" }
 attributes #4 = { convergent mustprogress noinline norecurse nounwind optnone "frame-pointer"="all" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" }
 attributes #5 = { convergent nounwind }
 ; no sycl-device-global-size attribute, this is not a device global variable but it contains compile-time properties,
 ; a metadata node will be generated.
-attributes #6 = { "sycl-unique-id"="6da74a122db9f35d____ZL7no_dg_int1" "device_image_scope"="false" "host_access"="1" "implement_in_csr"="true" "init_mode"="0" }
+attributes #6 = { "sycl-unique-id"="6da74a122db9f35d____ZL7no_dg_int1" "sycl-device-image-scope"="false" "sycl-host-access"="1" "sycl-implement-in-csr"="true" "sycl-init-mode"="0" }
 
 !llvm.dependent-libraries = !{!0}
 !llvm.module.flags = !{!1, !2}

llvm/test/tools/sycl-post-link/device-globals/test_global_variable_many_modules_no_dev_img_scope.ll

@@ -103,7 +103,7 @@ entry:
 }
 
 ; This device_global variable has no "device_image_scope" property
-attributes #0 = { "sycl-unique-id"="dg_int2" "host_access"="1" "implement_in_csr"="true" "init_mode"="0" "sycl-device-global-size"="4" }
+attributes #0 = { "sycl-unique-id"="dg_int2" "sycl-host-access"="1" "sycl-implement-in-csr"="true" "sycl-init-mode"="0" "sycl-device-global-size"="4" }
 attributes #1 = { convergent mustprogress noinline norecurse nounwind optnone "frame-pointer"="all" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" }
 attributes #2 = { convergent mustprogress noinline norecurse optnone "frame-pointer"="all" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="test_global_variable_1.cpp" "uniform-work-group-size"="true" }
 attributes #3 = { convergent mustprogress noinline norecurse optnone "frame-pointer"="all" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="test_global_variable_2.cpp" "uniform-work-group-size"="true" }

llvm/test/tools/sycl-post-link/device-globals/test_global_variable_many_modules_one_var_error.ll

@@ -112,7 +112,7 @@ entry:
   ret i32 addrspace(4)* %val
 }
 
-attributes #0 = { "sycl-unique-id"="dg_int2" "device_image_scope"="true" "host_access"="1" "implement_in_csr"="true" "init_mode"="0" "sycl-device-global-size"="4" }
+attributes #0 = { "sycl-unique-id"="dg_int2" "sycl-device-image-scope"="true" "sycl-host-access"="1" "sycl-implement-in-csr"="true" "sycl-init-mode"="0" "sycl-device-global-size"="4" }
 attributes #1 = { convergent mustprogress noinline norecurse nounwind optnone "frame-pointer"="all" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" }
 attributes #2 = { convergent mustprogress noinline norecurse optnone "frame-pointer"="all" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="test_global_variable_1.cpp" "uniform-work-group-size"="true" }
 attributes #3 = { convergent mustprogress noinline norecurse optnone "frame-pointer"="all" "min-legal-vector-width"="0" "no-trapping-math"="true" "stack-protector-buffer-size"="8" "sycl-module-id"="test_global_variable_2.cpp" "uniform-work-group-size"="true" }

llvm/tools/sycl-post-link/CompileTimeProperties.def

@@ -10,8 +10,9 @@
 #error "SYCL_COMPILE_TIME_PROPERTY(PropertyName, Decoration, ValueType) is not defined."
 #endif
 
-// The corresponding SPIR-V OpCodes for the init_mode and implement_in_csr
-// properties are documented in the SPV_INTEL_global_variable_decorations design
-// document: https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/DeviceGlobal/SPV_INTEL_global_variable_decorations.asciidoc#decoration
-SYCL_COMPILE_TIME_PROPERTY("init_mode", 6148, DecorValueTy::uint32)
-SYCL_COMPILE_TIME_PROPERTY("implement_in_csr", 6149, DecorValueTy::boolean)
+// The corresponding SPIR-V OpCodes for the sycl-init-mode and
+// sycl-implement-in-csr properties are documented in the
+// SPV_INTEL_global_variable_decorations design document:
+// https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/DeviceGlobal/SPV_INTEL_global_variable_decorations.asciidoc#decoration
+SYCL_COMPILE_TIME_PROPERTY("sycl-init-mode", 6148, DecorValueTy::uint32)
+SYCL_COMPILE_TIME_PROPERTY("sycl-implement-in-csr", 6149, DecorValueTy::boolean)

llvm/tools/sycl-post-link/CompileTimePropertiesPass.cpp

@@ -20,7 +20,7 @@ using namespace llvm;
 
 namespace {
 
-constexpr StringRef SYCL_HOST_ACCESS_ATTR = "host_access";
+constexpr StringRef SYCL_HOST_ACCESS_ATTR = "sycl-host-access";
 
 constexpr StringRef SPIRV_DECOR_MD_KIND = "spirv.Decorations";
 // The corresponding SPIR-V OpCode for the host_access property is documented

llvm/tools/sycl-post-link/DeviceGlobals.cpp

@@ -23,7 +23,7 @@ namespace {
 
 constexpr StringRef SYCL_DEVICE_GLOBAL_SIZE_ATTR = "sycl-device-global-size";
 constexpr StringRef SYCL_UNIQUE_ID_ATTR = "sycl-unique-id";
-constexpr StringRef SYCL_DEVICE_IMAGE_SCOPE_ATTR = "device_image_scope";
+constexpr StringRef SYCL_DEVICE_IMAGE_SCOPE_ATTR = "sycl-device-image-scope";
 
 /// Returns the size (in bytes) of the underlying type \c T of the device
 /// global variable.