diff --git a/src/compiler/glsl/ast.h b/src/compiler/glsl/ast.h index 10b9cf2608d..27ae471fadc 100644 --- a/src/compiler/glsl/ast.h +++ b/src/compiler/glsl/ast.h @@ -1353,14 +1353,14 @@ private: /** - * AST node representing a decalaration of the input layout for compute - * shaders. + * AST node representing a decalaration of the input layout for compute, + * task and mesh shaders. */ -class ast_cs_input_layout : public ast_node +class ast_cs_ms_input_layout : public ast_node { public: - ast_cs_input_layout(const struct YYLTYPE &locp, - ast_layout_expression *const *local_size) + ast_cs_ms_input_layout(const struct YYLTYPE &locp, + ast_layout_expression *const *local_size) { for (int i = 0; i < 3; i++) { this->local_size[i] = local_size[i]; diff --git a/src/compiler/glsl/ast_to_hir.cpp b/src/compiler/glsl/ast_to_hir.cpp index 1cede91886d..bf75be440b8 100644 --- a/src/compiler/glsl/ast_to_hir.cpp +++ b/src/compiler/glsl/ast_to_hir.cpp @@ -139,7 +139,7 @@ _mesa_ast_to_hir(ir_exec_list *instructions, struct _mesa_glsl_parse_state *stat state->gs_input_prim_type_specified = false; state->tcs_output_vertices_specified = false; - state->cs_input_local_size_specified = false; + state->cs_ms_input_local_size_specified = false; /* Section 4.2 of the GLSL 1.20 specification states: * "The built-in functions are scoped in a scope outside the global scope @@ -9002,8 +9002,8 @@ ast_gs_input_layout::hir(ir_exec_list *instructions, ir_rvalue * -ast_cs_input_layout::hir(ir_exec_list *instructions, - struct _mesa_glsl_parse_state *state) +ast_cs_ms_input_layout::hir(ir_exec_list *instructions, + struct _mesa_glsl_parse_state *state) { YYLTYPE loc = this->get_location(); @@ -9018,6 +9018,26 @@ ast_cs_input_layout::hir(ir_exec_list *instructions, * MAX_COMPUTE_WORK_GROUP_INVOCATIONS, but it seems reasonable to * report it at compile time as well. */ + + const unsigned *max_work_group_size; + unsigned max_work_group_invocations; + switch (state->stage) { + case MESA_SHADER_COMPUTE: + max_work_group_size = state->consts->MaxComputeWorkGroupSize; + max_work_group_invocations = state->consts->MaxComputeWorkGroupInvocations; + break; + case MESA_SHADER_TASK: + max_work_group_size = state->caps->mesh.max_task_work_group_size; + max_work_group_invocations = state->caps->mesh.max_task_work_group_invocations; + break; + case MESA_SHADER_MESH: + max_work_group_size = state->caps->mesh.max_mesh_work_group_size; + max_work_group_invocations = state->caps->mesh.max_mesh_work_group_invocations; + break; + default: + UNREACHABLE("invalid shader stage"); + } + GLuint64 total_invocations = 1; unsigned qual_local_size[3]; for (int i = 0; i < 3; i++) { @@ -9035,20 +9055,19 @@ ast_cs_input_layout::hir(ir_exec_list *instructions, } ralloc_free(local_size_str); - if (qual_local_size[i] > state->consts->MaxComputeWorkGroupSize[i]) { + if (qual_local_size[i] > max_work_group_size[i]) { _mesa_glsl_error(&loc, state, - "local_size_%c exceeds MAX_COMPUTE_WORK_GROUP_SIZE" + "local_size_%c exceeds max work group size" " (%d)", 'x' + i, - state->consts->MaxComputeWorkGroupSize[i]); + max_work_group_size[i]); break; } total_invocations *= qual_local_size[i]; - if (total_invocations > - state->consts->MaxComputeWorkGroupInvocations) { + if (total_invocations > max_work_group_invocations) { _mesa_glsl_error(&loc, state, "product of local_sizes exceeds " - "MAX_COMPUTE_WORK_GROUP_INVOCATIONS (%d)", - state->consts->MaxComputeWorkGroupInvocations); + "max work group invocations (%d)", + max_work_group_invocations); break; } } @@ -9056,11 +9075,11 @@ ast_cs_input_layout::hir(ir_exec_list *instructions, /* If any compute input layout declaration preceded this one, make sure it * was consistent with this one. */ - if (state->cs_input_local_size_specified) { + if (state->cs_ms_input_local_size_specified) { for (int i = 0; i < 3; i++) { - if (state->cs_input_local_size[i] != qual_local_size[i]) { + if (state->cs_ms_input_local_size[i] != qual_local_size[i]) { _mesa_glsl_error(&loc, state, - "compute shader input layout does not match" + "shader input layout does not match" " previous declaration"); return NULL; } @@ -9081,9 +9100,9 @@ ast_cs_input_layout::hir(ir_exec_list *instructions, return NULL; } - state->cs_input_local_size_specified = true; + state->cs_ms_input_local_size_specified = true; for (int i = 0; i < 3; i++) - state->cs_input_local_size[i] = qual_local_size[i]; + state->cs_ms_input_local_size[i] = qual_local_size[i]; /* We may now declare the built-in constant gl_WorkGroupSize (see * builtin_variable_generator::generate_constants() for why we didn't diff --git a/src/compiler/glsl/ast_type.cpp b/src/compiler/glsl/ast_type.cpp index 6c432f661f3..23a6d60e142 100644 --- a/src/compiler/glsl/ast_type.cpp +++ b/src/compiler/glsl/ast_type.cpp @@ -732,11 +732,16 @@ ast_type_qualifier::validate_in_qualifier(YYLTYPE *loc, valid_in_mask.flags.q.local_size_variable = 1; valid_in_mask.flags.q.derivative_group = 1; break; + case MESA_SHADER_TASK: + case MESA_SHADER_MESH: + valid_in_mask.flags.q.local_size = 7; + break; default: r = false; _mesa_glsl_error(loc, state, "input layout qualifiers only valid in " - "geometry, tessellation, fragment and compute shaders"); + "geometry, tessellation, fragment, task, mesh " + "and compute shaders"); break; } @@ -847,8 +852,8 @@ ast_type_qualifier::merge_into_in_qualifier(YYLTYPE *loc, * into HIR. */ if (state->in_qualifier->flags.q.local_size) { - node = new(lin_ctx) ast_cs_input_layout(*loc, - state->in_qualifier->local_size); + node = new(lin_ctx) ast_cs_ms_input_layout(*loc, + state->in_qualifier->local_size); state->in_qualifier->flags.q.local_size = 0; for (int i = 0; i < 3; i++) state->in_qualifier->local_size[i] = NULL; diff --git a/src/compiler/glsl/builtin_variables.cpp b/src/compiler/glsl/builtin_variables.cpp index d07bcbd9fca..4803fdfad3d 100644 --- a/src/compiler/glsl/builtin_variables.cpp +++ b/src/compiler/glsl/builtin_variables.cpp @@ -941,7 +941,7 @@ builtin_variable_generator::generate_constants() * * To prevent the shader from trying to refer to gl_WorkGroupSize before * the layout declaration, we don't define it here. Intead we define it - * in ast_cs_input_layout::hir(). + * in ast_cs_ms_input_layout::hir(). */ } diff --git a/src/compiler/glsl/gl_nir_linker.c b/src/compiler/glsl/gl_nir_linker.c index 134db1bf108..6a5296391b5 100644 --- a/src/compiler/glsl/gl_nir_linker.c +++ b/src/compiler/glsl/gl_nir_linker.c @@ -2564,6 +2564,41 @@ link_ms_inout_layout_qualifiers(struct gl_shader_program *prog, struct gl_shader **shader_list, unsigned num_shaders) { + /* handle task and mesh shader in layout */ + if (gl_prog->info.stage != MESA_SHADER_MESH && + gl_prog->info.stage != MESA_SHADER_TASK) + return; + + for (int i = 0; i < 3; i++) + gl_prog->nir->info.workgroup_size[i] = 0; + + for (unsigned sh = 0; sh < num_shaders; sh++) { + struct gl_shader *shader = shader_list[sh]; + + if (shader->info.Mesh.LocalSize[0] != 0) { + if (gl_prog->nir->info.workgroup_size[0] != 0) { + for (int i = 0; i < 3; i++) { + if (gl_prog->nir->info.workgroup_size[i] != + shader->info.Mesh.LocalSize[i]) { + linker_error(prog, "%s shader defined with conflicting local sizes\n", + gl_prog->info.stage == MESA_SHADER_TASK ? "task" : "mesh"); + return; + } + } + } + for (int i = 0; i < 3; i++) { + gl_prog->nir->info.workgroup_size[i] = shader->info.Mesh.LocalSize[i]; + } + } + } + + if (gl_prog->nir->info.workgroup_size[0] == 0) { + linker_error(prog, "%s shader must contain local group size\n", + gl_prog->info.stage == MESA_SHADER_TASK ? "task" : "mesh"); + return; + } + + /* handle mesh shader out layout */ if (gl_prog->info.stage != MESA_SHADER_MESH) return; diff --git a/src/compiler/glsl/glsl_parser.yy b/src/compiler/glsl/glsl_parser.yy index 812571d2e42..3a39836e459 100644 --- a/src/compiler/glsl/glsl_parser.yy +++ b/src/compiler/glsl/glsl_parser.yy @@ -1867,10 +1867,10 @@ layout_qualifier_id: for (int i = 0; i < 3; i++) { if (match_layout_qualifier(local_size_qualifiers[i], $1, state) == 0) { - if (!state->has_compute_shader()) { + if (!state->has_compute_shader() && !state->EXT_mesh_shader_enable) { _mesa_glsl_error(& @3, state, "%s qualifier requires GLSL 4.30 or " - "GLSL ES 3.10 or ARB_compute_shader", + "GLSL ES 3.10 or ARB_compute_shader or EXT_mesh_shader", local_size_qualifiers[i]); YYERROR; } else { diff --git a/src/compiler/glsl/glsl_parser_extras.cpp b/src/compiler/glsl/glsl_parser_extras.cpp index 1c3878db35d..81692174284 100644 --- a/src/compiler/glsl/glsl_parser_extras.cpp +++ b/src/compiler/glsl/glsl_parser_extras.cpp @@ -64,7 +64,7 @@ _mesa_glsl_parse_state::_mesa_glsl_parse_state(struct gl_context *_ctx, mesa_shader_stage stage, void *mem_ctx) : ctx(_ctx), exts(&_ctx->Extensions), consts(&_ctx->Const), caps(&_ctx->screen->caps), - api(_ctx->API), cs_input_local_size_specified(false), cs_input_local_size(), + api(_ctx->API), cs_ms_input_local_size_specified(false), cs_ms_input_local_size(), switch_state(), warnings_enabled(true) { assert(stage < MESA_SHADER_MESH_STAGES); @@ -1904,13 +1904,20 @@ set_shader_inout_layout(struct gl_shader *shader, /* Should have been prevented by the parser. */ if (shader->Stage != MESA_SHADER_GEOMETRY && shader->Stage != MESA_SHADER_TESS_EVAL && - shader->Stage != MESA_SHADER_COMPUTE) { + shader->Stage != MESA_SHADER_COMPUTE && + shader->Stage != MESA_SHADER_TASK && + shader->Stage != MESA_SHADER_MESH) { assert(!state->in_qualifier->flags.i); } + if (shader->Stage != MESA_SHADER_COMPUTE && + shader->Stage != MESA_SHADER_TASK && + shader->Stage != MESA_SHADER_MESH) { + assert(!state->cs_ms_input_local_size_specified); + } + if (shader->Stage != MESA_SHADER_COMPUTE) { /* Should have been prevented by the parser. */ - assert(!state->cs_input_local_size_specified); assert(!state->cs_input_local_size_variable_specified); assert(state->cs_derivative_group == DERIVATIVE_GROUP_NONE); } @@ -2040,9 +2047,9 @@ set_shader_inout_layout(struct gl_shader *shader, break; case MESA_SHADER_COMPUTE: - if (state->cs_input_local_size_specified) { + if (state->cs_ms_input_local_size_specified) { for (int i = 0; i < 3; i++) - shader->info.Comp.LocalSize[i] = state->cs_input_local_size[i]; + shader->info.Comp.LocalSize[i] = state->cs_ms_input_local_size[i]; } else { for (int i = 0; i < 3; i++) shader->info.Comp.LocalSize[i] = 0; @@ -2142,6 +2149,16 @@ set_shader_inout_layout(struct gl_shader *shader, shader->info.Mesh.MaxPrimitives = qual_max_primitives; } } + + FALLTHROUGH; + case MESA_SHADER_TASK: + if (state->cs_ms_input_local_size_specified) { + for (int i = 0; i < 3; i++) + shader->info.Mesh.LocalSize[i] = state->cs_ms_input_local_size[i]; + } else { + for (int i = 0; i < 3; i++) + shader->info.Mesh.LocalSize[i] = 0; + } break; default: diff --git a/src/compiler/glsl/glsl_parser_extras.h b/src/compiler/glsl/glsl_parser_extras.h index 7fef4b4648b..01bc81e0492 100644 --- a/src/compiler/glsl/glsl_parser_extras.h +++ b/src/compiler/glsl/glsl_parser_extras.h @@ -450,19 +450,19 @@ struct _mesa_glsl_parse_state { struct ast_type_qualifier *in_qualifier; /** - * True if a compute shader input local size was specified using a layout - * directive. + * True if a compute, task or mesh shader input local size was specified + * using a layout directive. * * Note: this value is computed at ast_to_hir time rather than at parse * time. */ - bool cs_input_local_size_specified; + bool cs_ms_input_local_size_specified; /** - * If cs_input_local_size_specified is true, the local size that was + * If cs_ms_input_local_size_specified is true, the local size that was * specified. Otherwise ignored. */ - unsigned cs_input_local_size[3]; + unsigned cs_ms_input_local_size[3]; /** * True if a compute shader input local variable size was specified using diff --git a/src/mesa/main/shader_types.h b/src/mesa/main/shader_types.h index f602074ae50..42aba9d1c7a 100644 --- a/src/mesa/main/shader_types.h +++ b/src/mesa/main/shader_types.h @@ -137,6 +137,8 @@ struct gl_shader_info * Mesh shader state from EXT_mesh_shader */ struct { + unsigned LocalSize[3]; + enum mesa_prim OutputType; GLint MaxVertices;