diff --git a/clang/include/clang/Basic/BuiltinsRISCV.def b/clang/include/clang/Basic/BuiltinsRISCV.def index a6a548f5c606..4d0e3bb148a4 100644 --- a/clang/include/clang/Basic/BuiltinsRISCV.def +++ b/clang/include/clang/Basic/BuiltinsRISCV.def @@ -18,7 +18,8 @@ //===----------------------------------------------------------------------===// // Ventus OpenCL builtins. //===----------------------------------------------------------------------===// - +// WORKAROUND: Disabled for now. +/* BUILTIN(__builtin_riscv_workgroup_id_x, "Ui", "nc") BUILTIN(__builtin_riscv_workgroup_id_y, "Ui", "nc") BUILTIN(__builtin_riscv_workgroup_id_z, "Ui", "nc") @@ -34,7 +35,7 @@ BUILTIN(__builtin_riscv_workgroup_size_z, "Us", "nc") BUILTIN(__builtin_riscv_grid_size_x, "Ui", "nc") BUILTIN(__builtin_riscv_grid_size_y, "Ui", "nc") BUILTIN(__builtin_riscv_grid_size_z, "Ui", "nc") - +*/ //===----------------------------------------------------------------------===// // Standard RISCV instruction builtins. //===----------------------------------------------------------------------===// diff --git a/clang/lib/CodeGen/CGBuiltin.cpp b/clang/lib/CodeGen/CGBuiltin.cpp index 08438d29c27e..7cb1db70c740 100644 --- a/clang/lib/CodeGen/CGBuiltin.cpp +++ b/clang/lib/CodeGen/CGBuiltin.cpp @@ -19461,6 +19461,8 @@ Value *CodeGenFunction::EmitRISCVBuiltinExpr(unsigned BuiltinID, switch (BuiltinID) { default: llvm_unreachable("unexpected builtin ID"); + // WORKAROUND: Disabled for now. + /* // Ventus GPGPU workitem case RISCV::BI__builtin_riscv_workitem_id_x: return emitRangedBuiltin(*this, Intrinsic::riscv_workitem_id_x, 0, 1024); @@ -19470,7 +19472,6 @@ Value *CodeGenFunction::EmitRISCVBuiltinExpr(unsigned BuiltinID, return emitRangedBuiltin(*this, Intrinsic::riscv_workitem_id_z, 0, 1024); // FIXME: Use AMDGPU function here temporarily - // Ventus GPGPU workgroup size case RISCV::BI__builtin_riscv_workgroup_size_x: return EmitAMDGPUWorkGroupSize(*this, 0); @@ -19486,7 +19487,7 @@ Value *CodeGenFunction::EmitRISCVBuiltinExpr(unsigned BuiltinID, return EmitAMDGPUGridSize(*this, 1); case RISCV::BI__builtin_riscv_grid_size_z: return EmitAMDGPUGridSize(*this, 2); - + */ case RISCV::BI__builtin_riscv_orc_b_32: case RISCV::BI__builtin_riscv_orc_b_64: case RISCV::BI__builtin_riscv_clz_32: