https://github.com/arsenm created https://github.com/llvm/llvm-project/pull/185206
These were just calling the __clc implementation, so move it to generic. >From 9dd021d271fd7727a7658e7bcfe6b34f21f7ca40 Mon Sep 17 00:00:00 2001 From: Matt Arsenault <[email protected]> Date: Sat, 7 Mar 2026 17:34:36 +0100 Subject: [PATCH] libclc: Remove target definitions of opencl workitem functions These were just calling the __clc implementation, so move it to generic. --- libclc/opencl/lib/amdgcn/SOURCES | 4 ---- libclc/opencl/lib/generic/SOURCES | 4 ++++ .../workitem/get_global_offset.cl | 0 .../{amdgcn => generic}/workitem/get_group_id.cl | 0 .../{amdgcn => generic}/workitem/get_local_id.cl | 0 .../{amdgcn => generic}/workitem/get_work_dim.cl | 0 libclc/opencl/lib/ptx-nvidiacl/SOURCES | 9 --------- .../lib/ptx-nvidiacl/workitem/get_global_id.cl | 13 ------------- .../lib/ptx-nvidiacl/workitem/get_group_id.cl | 13 ------------- .../lib/ptx-nvidiacl/workitem/get_local_id.cl | 13 ------------- .../ptx-nvidiacl/workitem/get_local_linear_id.cl | 13 ------------- .../ptx-nvidiacl/workitem/get_max_sub_group_size.cl | 13 ------------- .../lib/ptx-nvidiacl/workitem/get_num_sub_groups.cl | 13 ------------- .../lib/ptx-nvidiacl/workitem/get_sub_group_id.cl | 13 ------------- .../ptx-nvidiacl/workitem/get_sub_group_local_id.cl | 13 ------------- .../lib/ptx-nvidiacl/workitem/get_sub_group_size.cl | 13 ------------- 16 files changed, 4 insertions(+), 130 deletions(-) rename libclc/opencl/lib/{amdgcn => generic}/workitem/get_global_offset.cl (100%) rename libclc/opencl/lib/{amdgcn => generic}/workitem/get_group_id.cl (100%) rename libclc/opencl/lib/{amdgcn => generic}/workitem/get_local_id.cl (100%) rename libclc/opencl/lib/{amdgcn => generic}/workitem/get_work_dim.cl (100%) delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_global_id.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_group_id.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_id.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_linear_id.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_max_sub_group_size.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_num_sub_groups.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_id.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_local_id.cl delete mode 100644 libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_size.cl diff --git a/libclc/opencl/lib/amdgcn/SOURCES b/libclc/opencl/lib/amdgcn/SOURCES index e52f54789bfab..ad2b4ca6589ed 100644 --- a/libclc/opencl/lib/amdgcn/SOURCES +++ b/libclc/opencl/lib/amdgcn/SOURCES @@ -1,7 +1,3 @@ mem_fence/fence.cl subgroup/subgroup.cl synchronization/sub_group_barrier.cl -workitem/get_global_offset.cl -workitem/get_group_id.cl -workitem/get_local_id.cl -workitem/get_work_dim.cl diff --git a/libclc/opencl/lib/generic/SOURCES b/libclc/opencl/lib/generic/SOURCES index 18dd6fd3a10c2..398e326ffe482 100644 --- a/libclc/opencl/lib/generic/SOURCES +++ b/libclc/opencl/lib/generic/SOURCES @@ -206,6 +206,10 @@ workitem/get_enqueued_local_size.cl workitem/get_global_id.cl workitem/get_global_linear_id.cl workitem/get_global_size.cl +workitem/get_local_id.cl workitem/get_local_linear_id.cl workitem/get_local_size.cl workitem/get_num_groups.cl +workitem/get_global_offset.cl +workitem/get_group_id.cl +workitem/get_work_dim.cl diff --git a/libclc/opencl/lib/amdgcn/workitem/get_global_offset.cl b/libclc/opencl/lib/generic/workitem/get_global_offset.cl similarity index 100% rename from libclc/opencl/lib/amdgcn/workitem/get_global_offset.cl rename to libclc/opencl/lib/generic/workitem/get_global_offset.cl diff --git a/libclc/opencl/lib/amdgcn/workitem/get_group_id.cl b/libclc/opencl/lib/generic/workitem/get_group_id.cl similarity index 100% rename from libclc/opencl/lib/amdgcn/workitem/get_group_id.cl rename to libclc/opencl/lib/generic/workitem/get_group_id.cl diff --git a/libclc/opencl/lib/amdgcn/workitem/get_local_id.cl b/libclc/opencl/lib/generic/workitem/get_local_id.cl similarity index 100% rename from libclc/opencl/lib/amdgcn/workitem/get_local_id.cl rename to libclc/opencl/lib/generic/workitem/get_local_id.cl diff --git a/libclc/opencl/lib/amdgcn/workitem/get_work_dim.cl b/libclc/opencl/lib/generic/workitem/get_work_dim.cl similarity index 100% rename from libclc/opencl/lib/amdgcn/workitem/get_work_dim.cl rename to libclc/opencl/lib/generic/workitem/get_work_dim.cl diff --git a/libclc/opencl/lib/ptx-nvidiacl/SOURCES b/libclc/opencl/lib/ptx-nvidiacl/SOURCES index 3ece564c9760e..f20917346a3bc 100644 --- a/libclc/opencl/lib/ptx-nvidiacl/SOURCES +++ b/libclc/opencl/lib/ptx-nvidiacl/SOURCES @@ -1,10 +1 @@ mem_fence/fence.cl -workitem/get_global_id.cl -workitem/get_group_id.cl -workitem/get_local_id.cl -workitem/get_local_linear_id.cl -workitem/get_max_sub_group_size.cl -workitem/get_num_sub_groups.cl -workitem/get_sub_group_id.cl -workitem/get_sub_group_local_id.cl -workitem/get_sub_group_size.cl diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_global_id.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_global_id.cl deleted file mode 100644 index c84991ade76fc..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_global_id.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_global_id.h> - -_CLC_OVERLOAD _CLC_DEF size_t get_global_id(uint dim) { - return __clc_get_global_id(dim); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_group_id.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_group_id.cl deleted file mode 100644 index 6fc9832058736..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_group_id.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_group_id.h> - -_CLC_OVERLOAD _CLC_DEF size_t get_group_id(uint dim) { - return __clc_get_group_id(dim); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_id.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_id.cl deleted file mode 100644 index a2b608baf7aeb..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_id.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_local_id.h> - -_CLC_OVERLOAD _CLC_DEF size_t get_local_id(uint dim) { - return __clc_get_local_id(dim); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_linear_id.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_linear_id.cl deleted file mode 100644 index ccb301960726a..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_linear_id.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_local_linear_id.h> - -_CLC_OVERLOAD _CLC_DEF size_t get_local_linear_id() { - return __clc_get_local_linear_id(); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_max_sub_group_size.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_max_sub_group_size.cl deleted file mode 100644 index aa06c9561404b..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_max_sub_group_size.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_max_sub_group_size.h> - -_CLC_OVERLOAD _CLC_DEF uint get_max_sub_group_size() { - return __clc_get_max_sub_group_size(); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_num_sub_groups.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_num_sub_groups.cl deleted file mode 100644 index e23d78dafa5e8..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_num_sub_groups.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_num_sub_groups.h> - -_CLC_OVERLOAD _CLC_DEF uint get_num_sub_groups() { - return __clc_get_num_sub_groups(); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_id.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_id.cl deleted file mode 100644 index deb6134cd0704..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_id.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_sub_group_id.h> - -_CLC_OVERLOAD _CLC_DEF uint get_sub_group_id() { - return __clc_get_sub_group_id(); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_local_id.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_local_id.cl deleted file mode 100644 index 37487fbf3ece8..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_local_id.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_sub_group_local_id.h> - -_CLC_OVERLOAD _CLC_DEF uint get_sub_group_local_id() { - return __clc_get_sub_group_local_id(); -} diff --git a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_size.cl b/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_size.cl deleted file mode 100644 index e6be6f75171d1..0000000000000 --- a/libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_size.cl +++ /dev/null @@ -1,13 +0,0 @@ -//===----------------------------------------------------------------------===// -// -// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions. -// See https://llvm.org/LICENSE.txt for license information. -// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception -// -//===----------------------------------------------------------------------===// - -#include <clc/workitem/clc_get_sub_group_size.h> - -_CLC_OVERLOAD _CLC_DEF uint get_sub_group_size() { - return __clc_get_sub_group_size(); -} _______________________________________________ cfe-commits mailing list [email protected] https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
