[libclc] 0b2aaac - libclc: Remove target definitions of opencl workitem functions (#185206)
via cfe-commits
cfe-commits at lists.llvm.org
Sat Mar 7 23:15:42 PST 2026
Author: Matt Arsenault
Date: 2026-03-08T08:15:38+01:00
New Revision: 0b2aaac04a04cd471e2aa635d3462d8c22aa5a04
URL: https://github.com/llvm/llvm-project/commit/0b2aaac04a04cd471e2aa635d3462d8c22aa5a04
DIFF: https://github.com/llvm/llvm-project/commit/0b2aaac04a04cd471e2aa635d3462d8c22aa5a04.diff
LOG: libclc: Remove target definitions of opencl workitem functions (#185206)
These were just calling the __clc implementation, so move it
to generic.
Added:
libclc/opencl/lib/generic/workitem/get_global_offset.cl
libclc/opencl/lib/generic/workitem/get_group_id.cl
libclc/opencl/lib/generic/workitem/get_local_id.cl
libclc/opencl/lib/generic/workitem/get_work_dim.cl
Modified:
libclc/opencl/lib/amdgcn/SOURCES
libclc/opencl/lib/generic/SOURCES
libclc/opencl/lib/ptx-nvidiacl/SOURCES
Removed:
libclc/opencl/lib/amdgcn/workitem/get_global_offset.cl
libclc/opencl/lib/amdgcn/workitem/get_group_id.cl
libclc/opencl/lib/amdgcn/workitem/get_local_id.cl
libclc/opencl/lib/amdgcn/workitem/get_work_dim.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_global_id.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_group_id.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_id.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_local_linear_id.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_max_sub_group_size.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_num_sub_groups.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_id.cl
libclc/opencl/lib/ptx-nvidiacl/workitem/get_sub_group_local_id.cl
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 c9d29fb2d1905..4153d964e118d 100644
--- a/libclc/opencl/lib/amdgcn/SOURCES
+++ b/libclc/opencl/lib/amdgcn/SOURCES
@@ -2,7 +2,3 @@ async/wait_group_events.cl
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();
-}
More information about the cfe-commits
mailing list