Skip to content

Commit 46160a5

Browse files
Automerge: [SPIR-V] Specify target environment in tests referring to the BuiltIn WorkgroupSize variable (#122755)
KhronosGroup/SPIRV-Tools#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.
2 parents 4c4adbd + b74d3e1 commit 46160a5

File tree

4 files changed

+6
-6
lines changed

4 files changed

+6
-6
lines changed

llvm/test/CodeGen/SPIRV/pointers/type-deduce-sycl-stub.ll

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1,8 +1,8 @@
11
; RUN: llc -O0 -mtriple=spirv64-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
2-
; 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 %}
2+
; 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 %}
33

44
; RUN: llc -O0 -mtriple=spirv32-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
5-
; 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 %}
5+
; 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 %}
66

77
; CHECK-SPIRV-DAG: OpName %[[#F:]] "finish"
88
; CHECK-SPIRV-DAG: OpName %[[#FH:]] "finish_helper"

llvm/test/CodeGen/SPIRV/transcoding/OpGroupBroadcast.ll

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -1,8 +1,8 @@
11
; RUN: llc -O0 -mtriple=spirv64-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
2-
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val %}
2+
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
33

44
; RUN: llc -O0 -mtriple=spirv32-unknown-unknown %s -o - | FileCheck %s --check-prefix=CHECK-SPIRV
5-
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv32-unknown-unknown %s -o - -filetype=obj | spirv-val %}
5+
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv32-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
66

77
; CHECK-SPIRV: OpCapability Groups
88
; CHECK-SPIRV-DAG: %[[#Int32Ty:]] = OpTypeInt 32 0

llvm/test/CodeGen/SPIRV/validate/sycl-hier-par-basic.ll

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -2,7 +2,7 @@
22
; that reproduced multiple cases of the issues when OpPhi's result type mismatches with operand types.
33
; The only pass criterion is that spirv-val considers output valid.
44

5-
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val %}
5+
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
66

77
%struct.PFWGFunctor = type { i64, i64, i32, i32, %"class.sycl::_V1::accessor" }
88
%"class.sycl::_V1::accessor" = type { %"class.sycl::_V1::detail::AccessorImplDevice", %union.anon }

llvm/test/CodeGen/SPIRV/validate/sycl-tangle-group-algorithms.ll

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -8,7 +8,7 @@
88

99
; The only pass criterion is that spirv-val considers output valid.
1010

11-
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64v1.5-unknown-unknown %s -o - -filetype=obj | spirv-val %}
11+
; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv64v1.5-unknown-unknown %s -o - -filetype=obj | spirv-val --target-env spv1.4 %}
1212

1313
%"nd_item" = type { i8 }
1414
%struct.AssertHappened = type { i32, [257 x i8], [257 x i8], [129 x i8], i32, i64, i64, i64, i64, i64, i64 }

0 commit comments

Comments
 (0)