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

[Vyacheslav Levytskyy via llvm-commits]

https://github.com/VyacheslavLevytskyy created https://github.com/llvm/llvm-project/pull/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.

>From 1c9c93be6591613a70043c3911aa63be2e03addb Mon Sep 17 00:00:00 2001
From: "Levytskyy, Vyacheslav" <vyacheslav.levytskyy at intel.com>
Date: Mon, 13 Jan 2025 09:53:59 -0800
Subject: [PATCH] Specify target environment in tests referring to the BuiltIn
 WorkgroupSize variable

---
 llvm/test/CodeGen/SPIRV/pointers/type-deduce-sycl-stub.ll     | 4 ++--
 llvm/test/CodeGen/SPIRV/transcoding/OpGroupBroadcast.ll       | 4 ++--
 llvm/test/CodeGen/SPIRV/validate/sycl-hier-par-basic.ll       | 2 +-
 .../CodeGen/SPIRV/validate/sycl-tangle-group-algorithms.ll    | 2 +-
 4 files changed, 6 insertions(+), 6 deletions(-)

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 }