Skip to content

[REFACTOR][TARGET] Cleanup backend target registration #19759

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
spectrometerHBH merged 1 commit into from
Jun 14, 2026
Merged

[REFACTOR][TARGET] Cleanup backend target registration #19759

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
1 commit
Select commit
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 1 addition & 1 deletion CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -349,7 +349,7 @@ tvm_file_glob(GLOB CODEGEN_SRCS
src/backend/trn/codegen/*.cc
src/backend/trn/op/*.cc
src/backend/trn/transform/*.cc
src/backend/vulkan/codegen/register.cc
src/backend/vulkan/codegen/target_kind.cc

@gemini-code-assist gemini-code-assist Bot Jun 14, 2026

Copy link
Copy Markdown
Contributor

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

high

The file src/target/llvm/intrin_rule_nvptx.cc has been renamed to src/backend/cuda/codegen/llvm/intrin_rule_nvptx.cc. However, because CMakeLists.txt only globs src/backend/cuda/codegen/*.cc (non-recursively), and cmake/modules/LLVM.cmake likely only globs src/target/llvm/*.cc, this file is now excluded from the build. To ensure the NVPTX intrinsic rules are compiled, we should add src/backend/cuda/codegen/llvm/*.cc to the CODEGEN_SRCS list.

  src/backend/cuda/codegen/llvm/*.cc
  src/backend/vulkan/codegen/target_kind.cc

src/backend/vulkan/codegen/vulkan_fallback_module.cc
src/backend/webgpu/codegen/*.cc
)
Expand Down
0 src/target/llvm/intrin_rule_nvptx.cc → ...nd/cuda/codegen/llvm/intrin_rule_nvptx.cc
View file
Open in desktop
File renamed without changes.
2 changes: 1 addition & 1 deletion src/backend/cuda/codegen/register.cc → src/backend/cuda/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief CUDA compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
35 changes: 0 additions & 35 deletions src/backend/cuda/op/register.cc
View file
Open in desktop

This file was deleted.

4 changes: 4 additions & 0 deletions src/backend/cuda/op/target_builtin.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,6 +22,8 @@
*
* builtin intrinsic operators specific to CUDA target.
*/
#include <tvm/ffi/function.h>
#include <tvm/runtime/base.h>
#include <tvm/tirx/builtin.h>
#include <tvm/tirx/op.h>
#include <tvm/tirx/op_attr_types.h>
Expand Down Expand Up @@ -597,6 +599,8 @@ void RegisterDeviceIntrinsicAliases() {

#undef TIRX_DEFINE_BUILTIN_FUNC

TVM_FFI_STATIC_INIT_BLOCK() { RegisterCudaTargetBuiltins(); }

} // namespace builtin
} // namespace tirx
} // namespace tvm
2 changes: 1 addition & 1 deletion src/backend/hexagon/codegen/register.cc → src/backend/hexagon/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief Hexagon compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
2 changes: 1 addition & 1 deletion src/backend/metal/codegen/register.cc → src/backend/metal/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief Metal compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
35 changes: 0 additions & 35 deletions src/backend/metal/op/register.cc
View file
Open in desktop

This file was deleted.

4 changes: 4 additions & 0 deletions src/backend/metal/op/target_builtin.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -22,6 +22,8 @@
*
* builtin intrinsic operators specific to Metal target.
*/
#include <tvm/ffi/function.h>
#include <tvm/runtime/base.h>
#include <tvm/tirx/op.h>
#include <tvm/tirx/op_attr_types.h>

Expand Down Expand Up @@ -57,6 +59,8 @@ TIRX_DEFINE_BUILTIN_FUNC(simdgroup_multiply_accumulate)

#undef TIRX_DEFINE_BUILTIN_FUNC

TVM_FFI_STATIC_INIT_BLOCK() { RegisterMetalTargetBuiltins(); }

} // namespace builtin
} // namespace tirx
} // namespace tvm
2 changes: 1 addition & 1 deletion src/backend/opencl/codegen/register.cc → src/backend/opencl/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief OpenCL compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
2 changes: 1 addition & 1 deletion src/backend/rocm/codegen/register.cc → src/backend/rocm/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief ROCm compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
2 changes: 1 addition & 1 deletion src/backend/trn/codegen/register.cc → src/backend/trn/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief Trainium compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
35 changes: 0 additions & 35 deletions src/backend/trn/op/register.cc
View file
Open in desktop

This file was deleted.

4 changes: 4 additions & 0 deletions src/backend/trn/op/target_builtin.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -23,6 +23,8 @@
*
* builtin intrinsic operators specific to Trainium target.
*/
#include <tvm/ffi/function.h>
#include <tvm/runtime/base.h>
#include <tvm/tirx/builtin.h>
#include <tvm/tirx/op.h>
#include <tvm/tirx/op_attr_types.h>
Expand Down Expand Up @@ -161,6 +163,8 @@ void RegisterNKIIntrinsicAliases() {

#undef TIRX_DEFINE_BUILTIN_FUNC

TVM_FFI_STATIC_INIT_BLOCK() { RegisterTRNTargetBuiltins(); }

} // namespace builtin
} // namespace tirx
} // namespace tvm
2 changes: 1 addition & 1 deletion src/backend/vulkan/codegen/register.cc → src/backend/vulkan/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief Vulkan compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
2 changes: 1 addition & 1 deletion src/backend/webgpu/codegen/register.cc → src/backend/webgpu/codegen/target_kind.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -18,7 +18,7 @@
*/

/*!
* \file register.cc
* \file target_kind.cc
* \brief WebGPU compiler backend static registration.
*/
#include <dlpack/dlpack.h>
Expand Down
1 change: 0 additions & 1 deletion src/target/opt/README
View file
Open in desktop

This file was deleted.

Loading