OpenHarmony/third_party_mesa3d
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

intel/compiler: Add a SIMD_COUNT constant

Browse Source 
Reviewed-by: Lionel Landwerlin <lionel.g.landwerlin@intel.com>
Reviewed-by: Ivan Briano <ivan.briano@intel.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/19601>
This commit is contained in:
Caio Oliveira
2022-11-08 02:30:08 -08:00
committed by Marge Bot
parent a0580dadfd
commit 3c52e2d04c
2 changed files with 13 additions and 4 deletions
Download Patch File Download Diff File Expand all files Collapse all files

5
src/intel/compiler/brw_private.h
View File

@@ -29,7 +29,10 @@
unsigned brw_required_dispatch_width(const struct shader_info *info); unsigned brw_required_dispatch_width(const struct shader_info *info);
static constexpr int SIMD_COUNT = 3;
struct brw_simd_selection_state { struct brw_simd_selection_state {
void *mem_ctx; void *mem_ctx;
const struct intel_device_info *devinfo; const struct intel_device_info *devinfo;
@@ -37,7 +40,7 @@ struct brw_simd_selection_state {
unsigned required_width; unsigned required_width;
const char *error[3]; const char *error[SIMD_COUNT];
}; };
bool brw_simd_should_compile(brw_simd_selection_state &state, unsigned simd); bool brw_simd_should_compile(brw_simd_selection_state &state, unsigned simd);

12
src/intel/compiler/brw_simd_selection.cpp
View File

@@ -50,6 +50,8 @@ bool
brw_simd_should_compile(brw_simd_selection_state &state, brw_simd_should_compile(brw_simd_selection_state &state,
unsigned simd) unsigned simd)
{ {
assert(simd < SIMD_COUNT);
struct brw_cs_prog_data *prog_data = state.prog_data; struct brw_cs_prog_data *prog_data = state.prog_data;
assert(!test_bit(prog_data->prog_mask, simd)); assert(!test_bit(prog_data->prog_mask, simd));
@@ -123,12 +125,14 @@ brw_simd_should_compile(brw_simd_selection_state &state,
return false; return false;
} }
const bool env_skip[3] = { static const bool env_skip[] = {
INTEL_DEBUG(DEBUG_NO8) != 0, INTEL_DEBUG(DEBUG_NO8) != 0,
INTEL_DEBUG(DEBUG_NO16) != 0, INTEL_DEBUG(DEBUG_NO16) != 0,
INTEL_DEBUG(DEBUG_NO32) != 0, INTEL_DEBUG(DEBUG_NO32) != 0,
}; };
static_assert(ARRAY_SIZE(env_skip) == SIMD_COUNT);
if (unlikely(env_skip[simd])) { if (unlikely(env_skip[simd])) {
state.error[simd] = ralloc_asprintf( state.error[simd] = ralloc_asprintf(
state.mem_ctx, "SIMD%u skipped because INTEL_DEBUG=no%u", state.mem_ctx, "SIMD%u skipped because INTEL_DEBUG=no%u",
@@ -142,6 +146,8 @@ brw_simd_should_compile(brw_simd_selection_state &state,
void void
brw_simd_mark_compiled(brw_simd_selection_state &state, unsigned simd, bool spilled) brw_simd_mark_compiled(brw_simd_selection_state &state, unsigned simd, bool spilled)
{ {
assert(simd < SIMD_COUNT);
struct brw_cs_prog_data *prog_data = state.prog_data; struct brw_cs_prog_data *prog_data = state.prog_data;
assert(!test_bit(prog_data->prog_mask, simd)); assert(!test_bit(prog_data->prog_mask, simd));
@@ -149,7 +155,7 @@ brw_simd_mark_compiled(brw_simd_selection_state &state, unsigned simd, bool spil
/* If a SIMD spilled, all the larger ones would spill too. */ /* If a SIMD spilled, all the larger ones would spill too. */
if (spilled) { if (spilled) {
for (unsigned i = simd; i < 3; i++) for (unsigned i = simd; i < SIMD_COUNT; i++)
prog_data->prog_spilled |= 1u << i; prog_data->prog_spilled |= 1u << i;
} }
} }
@@ -201,7 +207,7 @@ brw_simd_select_for_workgroup_size(const struct intel_device_info *devinfo,
.prog_data = &cloned, .prog_data = &cloned,
}; };
for (unsigned simd = 0; simd < 3; simd++) { for (unsigned simd = 0; simd < SIMD_COUNT; simd++) {
/* We are not recompiling, so use original results of prog_mask and /* We are not recompiling, so use original results of prog_mask and
* prog_spilled as they will already contain all possible compilations. * prog_spilled as they will already contain all possible compilations.
*/ */