diff --git a/src/intel/compiler/brw/brw_shader.cpp b/src/intel/compiler/brw/brw_shader.cpp index e0f91b81dba..3f85615413d 100644 --- a/src/intel/compiler/brw/brw_shader.cpp +++ b/src/intel/compiler/brw/brw_shader.cpp @@ -17,6 +17,7 @@ #include "dev/intel_wa.h" #include "compiler/glsl_types.h" #include "compiler/nir/nir_builder.h" +#include "util/bitscan.h" #include "util/u_math.h" void @@ -999,7 +1000,17 @@ brw_cs_get_dispatch_info(const struct intel_device_info *devinfo, override_local_size ? override_local_size : prog_data->local_size; - const int simd = brw_simd_select_for_workgroup_size(devinfo, prog_data, sizes); + int simd = -1; + if (intel_use_jay(devinfo, MESA_SHADER_COMPUTE)) { + /* Currently Jay compiles only a single binary, just select that. In the + * future this needs to get smarter. + */ + assert(util_is_power_of_two_nonzero(prog_data->prog_mask)); + simd = util_logbase2(prog_data->prog_mask); + } else { + simd = brw_simd_select_for_workgroup_size(devinfo, prog_data, sizes); + } + assert(simd >= 0 && simd < 3); info.group_size = sizes[0] * sizes[1] * sizes[2];