From 202fe3de31c8db5d62ed05e3db6759aa3c6f5d7d Mon Sep 17 00:00:00 2001 From: Karol Herbst Date: Sat, 26 Aug 2023 15:54:47 +0200 Subject: [PATCH] intel/compiler: drop 64 bit handling for cl workgroup intrinsics Signed-off-by: Karol Herbst Reviewed-by: Alyssa Rosenzweig Part-of: --- src/intel/compiler/brw_kernel.c | 7 +------ src/intel/compiler/brw_nir_lower_cs_intrinsics.c | 13 ------------- 2 files changed, 1 insertion(+), 19 deletions(-) diff --git a/src/intel/compiler/brw_kernel.c b/src/intel/compiler/brw_kernel.c index ea0aa2d248c..8d00b2a706d 100644 --- a/src/intel/compiler/brw_kernel.c +++ b/src/intel/compiler/brw_kernel.c @@ -214,12 +214,7 @@ lower_kernel_intrinsics(nir_shader *nir) nir_intrinsic_set_range(load, 3 * 4); nir_def_init(&load->instr, &load->def, 3, 32); nir_builder_instr_insert(&b, &load->instr); - - /* We may need to do a bit-size cast here */ - nir_def *num_work_groups = - nir_u2uN(&b, &load->def, intrin->def.bit_size); - - nir_def_rewrite_uses(&intrin->def, num_work_groups); + nir_def_rewrite_uses(&intrin->def, &load->def); progress = true; break; } diff --git a/src/intel/compiler/brw_nir_lower_cs_intrinsics.c b/src/intel/compiler/brw_nir_lower_cs_intrinsics.c index fabe16858d2..9088194986f 100644 --- a/src/intel/compiler/brw_nir_lower_cs_intrinsics.c +++ b/src/intel/compiler/brw_nir_lower_cs_intrinsics.c @@ -189,19 +189,6 @@ lower_cs_intrinsics_convert_block(struct lower_intrinsics_state *state, nir_def *sysval; switch (intrinsic->intrinsic) { - case nir_intrinsic_load_workgroup_size: - case nir_intrinsic_load_workgroup_id: - case nir_intrinsic_load_num_workgroups: - /* Convert this to 32-bit if it's not */ - if (intrinsic->def.bit_size == 64) { - intrinsic->def.bit_size = 32; - sysval = nir_u2u64(b, &intrinsic->def); - nir_def_rewrite_uses_after(&intrinsic->def, - sysval, - sysval->parent_instr); - } - continue; - case nir_intrinsic_load_local_invocation_index: case nir_intrinsic_load_local_invocation_id: { if (!local_index && !nir->info.workgroup_size_variable) {