[llvm] b74d3e1 - [SPIR-V] Specify target environment in tests referring to the BuiltIn WorkgroupSize variable (#122755)

[via llvm-commits]

Author: Vyacheslav Levytskyy
Date: 2025-01-14T17:44:07+01:00
New Revision: b74d3e179d6d1d8aad65a7ee8d359defd94a8ec1

URL: https://github.com/llvm/llvm-project/commit/b74d3e179d6d1d8aad65a7ee8d359defd94a8ec1
DIFF: https://github.com/llvm/llvm-project/commit/b74d3e179d6d1d8aad65a7ee8d359defd94a8ec1.diff

LOG: [SPIR-V] Specify target environment in tests referring to the BuiltIn WorkgroupSize variable (#122755)

https://github.com/KhronosGroup/SPIRV-Tools/pull/5407 introduces a check
for WorkgroupSize variable to be a 3-component 32-bit int vector, and
indeed, we see this requirement in
https://registry.khronos.org/vulkan/specs/latest/man/html/WorkgroupSize.html#VUID-WorkgroupSize-WorkgroupSize-04427

However, OpenCL imposes different requirements, documented here:
https://registry.khronos.org/OpenCL/specs/3.0-unified/html/OpenCL_Env.html#_built_in_variables
OpenCL environment requires WorkgroupSize variable to have components of
size_t size that will be 32 or 64 depending on a target. This is the way
how the SPIR-V Backend implements this, by querying pointer size of the
current platform/target.

To allow spirv-val to account target environments difference, this PR
adds `--target-env <env>` to test cases referring to the BuiltIn
WorkgroupSize variable.

Added: 
    

Modified: 
    llvm/test/CodeGen/SPIRV/pointers/type-deduce-sycl-stub.ll
    llvm/test/CodeGen/SPIRV/transcoding/OpGroupBroadcast.ll
    llvm/test/CodeGen/SPIRV/validate/sycl-hier-par-basic.ll
    llvm/test/CodeGen/SPIRV/validate/sycl-tangle-group-algorithms.ll

Removed: 
    


################################################################################
diff  --git a/llvm/test/CodeGen/SPIRV/pointers/type-deduce-sycl-stub.ll b/llvm/test/CodeGen/SPIRV/pointers/type-deduce-sycl-stub.ll
index 008a474f1c104b..ab3655df2200f8 100644
--- a/llvm/test/CodeGen/SPIRV/pointers/type-deduce-sycl-stub.ll
+++ b/llvm/test/CodeGen/SPIRV/pointers/type-deduce-sycl-stub.ll
@@ -1,8 +1,8 @@
 ; RUN: llc -O0 -mtriple=spirv64-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
-; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown --spirv-ext=all --translator-compatibility-mode --avoid-spirv-capabilities=Shader %s -o - -filetype=obj | spirv-val %}
+; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown --spirv-ext=all --translator-compatibility-mode --avoid-spirv-capabilities=Shader %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
 
 ; RUN: llc -O0 -mtriple=spirv32-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
-; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv32-unknown-unknown --spirv-ext=all --translator-compatibility-mode --avoid-spirv-capabilities=Shader %s -o - -filetype=obj | spirv-val %}
+; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv32-unknown-unknown --spirv-ext=all --translator-compatibility-mode --avoid-spirv-capabilities=Shader %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
 
 ; CHECK-SPIRV-DAG: OpName %[[#F:]] "finish"
 ; CHECK-SPIRV-DAG: OpName %[[#FH:]] "finish_helper"

diff  --git a/llvm/test/CodeGen/SPIRV/transcoding/OpGroupBroadcast.ll b/llvm/test/CodeGen/SPIRV/transcoding/OpGroupBroadcast.ll
index fd737e85a8ec50..c60f70a95d457c 100644
--- a/llvm/test/CodeGen/SPIRV/transcoding/OpGroupBroadcast.ll
+++ b/llvm/test/CodeGen/SPIRV/transcoding/OpGroupBroadcast.ll
@@ -1,8 +1,8 @@
 ; RUN: llc -O0 -mtriple=spirv64-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
-; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val %}
+; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
 
 ; RUN: llc -O0 -mtriple=spirv32-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
-; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv32-unknown-unknown %s -o - -filetype=obj | spirv-val %}
+; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv32-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
 
 ; CHECK-SPIRV: OpCapability Groups
 ; CHECK-SPIRV-DAG: %[[#Int32Ty:]] = OpTypeInt 32 0

diff  --git a/llvm/test/CodeGen/SPIRV/validate/sycl-hier-par-basic.ll b/llvm/test/CodeGen/SPIRV/validate/sycl-hier-par-basic.ll
index 77ed1d6fecf9ae..00ed559dac0964 100644
--- a/llvm/test/CodeGen/SPIRV/validate/sycl-hier-par-basic.ll
+++ b/llvm/test/CodeGen/SPIRV/validate/sycl-hier-par-basic.ll
@@ -2,7 +2,7 @@
 ; that reproduced multiple cases of the issues when OpPhi's result type mismatches with operand types.
 ; The only pass criterion is that spirv-val considers output valid.
 
-; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val %}
+; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
 
 %struct.PFWGFunctor = type { i64, i64, i32, i32, %"class.sycl::_V1::accessor" }
 %"class.sycl::_V1::accessor" = type { %"class.sycl::_V1::detail::AccessorImplDevice", %union.anon }

diff  --git a/llvm/test/CodeGen/SPIRV/validate/sycl-tangle-group-algorithms.ll b/llvm/test/CodeGen/SPIRV/validate/sycl-tangle-group-algorithms.ll
index b6b919f36d92c6..a10341bce4859a 100644
--- a/llvm/test/CodeGen/SPIRV/validate/sycl-tangle-group-algorithms.ll
+++ b/llvm/test/CodeGen/SPIRV/validate/sycl-tangle-group-algorithms.ll
@@ -8,7 +8,7 @@
 
 ; The only pass criterion is that spirv-val considers output valid.
 
-; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64v1.5-unknown-unknown %s -o - -filetype=obj | spirv-val %}
+; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64v1.5-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
 
 %"nd_item" = type { i8 }
 %struct.AssertHappened = type { i32, [257 x i8], [257 x i8], [129 x i8], i32, i64, i64, i64, i64, i64, i64 }