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

zink: add use_local_size flag to compute programs

Browse Source 
Reviewed-by: Dave Airlie <airlied@redhat.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/18197>
This commit is contained in:
Mike Blumenkrantz
2022-08-12 10:23:09 -04:00
committed by Marge Bot
parent bd12fa86fc
commit 3a674fc6eb
2 changed files with 8 additions and 6 deletions
Download Patch File Download Diff File Expand all files Collapse all files

12
src/gallium/drivers/zink/zink_program.c
View File

@@ -698,13 +698,9 @@ hash_compute_pipeline_state(const void *key)
void
zink_program_update_compute_pipeline_state(struct zink_context *ctx, struct zink_compute_program *comp, const uint block[3])
{
struct zink_shader *zs = comp->shader;
bool use_local_size = !(zs->nir->info.workgroup_size[0] ||
zs->nir->info.workgroup_size[1] ||
zs->nir->info.workgroup_size[2]);
if (ctx->compute_pipeline_state.use_local_size != use_local_size)
if (ctx->compute_pipeline_state.use_local_size != comp->use_local_size)
ctx->compute_pipeline_state.dirty = true;
ctx->compute_pipeline_state.use_local_size = use_local_size;
ctx->compute_pipeline_state.use_local_size = comp->use_local_size;
if (ctx->compute_pipeline_state.use_local_size) {
for (int i = 0; i < ARRAY_SIZE(ctx->compute_pipeline_state.local_size); i++) {
@@ -743,6 +739,10 @@ create_compute_program(struct zink_context *ctx, nir_shader *nir)
util_dynarray_init(&comp->shader_cache[0], NULL);
util_dynarray_init(&comp->shader_cache[1], NULL);
comp->use_local_size = !(nir->info.workgroup_size[0] ||
nir->info.workgroup_size[1] ||
nir->info.workgroup_size[2]);
comp->pipelines = _mesa_hash_table_create(NULL, NULL,
equals_compute_pipeline_state);

2
src/gallium/drivers/zink/zink_types.h
View File

@@ -803,6 +803,8 @@ struct zink_gfx_program {
struct zink_compute_program {
struct zink_program base;
bool use_local_size;
struct zink_shader_module *curr;
struct zink_shader_module *module; //base