fdo-mirrors/mesa
1
0
Fork 0
mirror of https://gitlab.freedesktop.org/mesa/mesa.git synced 2026-01-01 16:10:09 +01:00
Code Issues Projects Releases Packages Wiki Activity Actions 3

i965/cs: Rework cs_emit to take a nir_shader and a brw_compiler

Browse source 
This commit removes all dependence on GL state by getting rid of the
brw_context parameter and the GL data structures.

Reviewed-by: Topi Pohjolainen <topi.pohjolainen@intel.com>
This commit is contained in:
Jason Ekstrand 2015-10-08 15:28:26 -07:00
parent 657863bb5c
commit 4e711872d0
3 changed files with 42 additions and 29 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

10
src/mesa/drivers/dri/i965/brw_cs.c
View file

@ -105,9 +105,15 @@ brw_codegen_cs_prog(struct brw_context *brw,
if (INTEL_DEBUG & DEBUG_SHADER_TIME)
st_index = brw_get_shader_time_index(brw, prog, &cp->program.Base, ST_CS);
program = brw_cs_emit(brw, mem_ctx, key, &prog_data,
&cp->program, prog, st_index, &program_size);
char *error_str;
program = brw_cs_emit(brw->intelScreen->compiler, brw, mem_ctx,
key, &prog_data, cp->program.Base.nir,
st_index, &program_size, &error_str);
if (program == NULL) {
prog->LinkStatus = false;
ralloc_strcat(&prog->InfoLog, error_str);
_mesa_problem(NULL, "Failed to compile compute shader: %s\n", error_str);
ralloc_free(mem_ctx);
return false;
}

10
src/mesa/drivers/dri/i965/brw_cs.h
View file

@ -39,15 +39,17 @@ extern "C" {
void
brw_upload_cs_prog(struct brw_context *brw);
struct nir_shader;
const unsigned *
brw_cs_emit(struct brw_context *brw,
brw_cs_emit(const struct brw_compiler *compiler, void *log_data,
void *mem_ctx,
const struct brw_cs_prog_key *key,
struct brw_cs_prog_data *prog_data,
struct gl_compute_program *cp,
struct gl_shader_program *prog,
const struct nir_shader *shader,
int shader_time_index,
unsigned *final_assembly_size);
unsigned *final_assembly_size,
char **error_str);
void
brw_cs_fill_local_id_payload(const struct brw_cs_prog_data *cs_prog_data,

51
src/mesa/drivers/dri/i965/brw_fs.cpp
View file

@ -5234,29 +5234,32 @@ fs_visitor::emit_cs_work_group_id_setup()
}
const unsigned *
brw_cs_emit(struct brw_context *brw,
brw_cs_emit(const struct brw_compiler *compiler, void *log_data,
void *mem_ctx,
const struct brw_cs_prog_key *key,
struct brw_cs_prog_data *prog_data,
struct gl_compute_program *cp,
struct gl_shader_program *prog,
const nir_shader *shader,
int shader_time_index,
unsigned *final_assembly_size)
unsigned *final_assembly_size,
char **error_str)
{
prog_data->local_size[0] = cp->LocalSize[0];
prog_data->local_size[1] = cp->LocalSize[1];
prog_data->local_size[2] = cp->LocalSize[2];
prog_data->local_size[0] = shader->info.cs.local_size[0];
prog_data->local_size[1] = shader->info.cs.local_size[1];
prog_data->local_size[2] = shader->info.cs.local_size[2];
unsigned local_workgroup_size =
cp->LocalSize[0] * cp->LocalSize[1] * cp->LocalSize[2];
unsigned max_cs_threads = brw->intelScreen->compiler->devinfo->max_cs_threads;
shader->info.cs.local_size[0] * shader->info.cs.local_size[1] *
shader->info.cs.local_size[2];
unsigned max_cs_threads = compiler->devinfo->max_cs_threads;
cfg_t *cfg = NULL;
const char *fail_msg = NULL;
/* Now the main event: Visit the shader IR and generate our CS IR for it.
*/
fs_visitor v8(brw->intelScreen->compiler, brw, mem_ctx, key,
&prog_data->base, &cp->Base, cp->Base.nir, 8, shader_time_index);
fs_visitor v8(compiler, log_data, mem_ctx, key, &prog_data->base,
NULL, /* Never used in core profile */
shader, 8, shader_time_index);
if (!v8.run_cs()) {
fail_msg = v8.fail_msg;
} else if (local_workgroup_size <= 8 * max_cs_threads) {
@ -5264,15 +5267,18 @@ brw_cs_emit(struct brw_context *brw,
prog_data->simd_size = 8;
}
fs_visitor v16(brw->intelScreen->compiler, brw, mem_ctx, key,
&prog_data->base, &cp->Base, cp->Base.nir, 16, shader_time_index);
fs_visitor v16(compiler, log_data, mem_ctx, key, &prog_data->base,
NULL, /* Never used in core profile */
shader, 16, shader_time_index);
if (likely(!(INTEL_DEBUG & DEBUG_NO16)) &&
!fail_msg && !v8.simd16_unsupported &&
local_workgroup_size <= 16 * max_cs_threads) {
/* Try a SIMD16 compile */
v16.import_uniforms(&v8);
if (!v16.run_cs()) {
perf_debug("SIMD16 shader failed to compile: %s", v16.fail_msg);
compiler->shader_perf_log(log_data,
"SIMD16 shader failed to compile: %s",
v16.fail_msg);
if (!cfg) {
fail_msg =
"Couldn't generate SIMD16 program and not "
@ -5286,20 +5292,19 @@ brw_cs_emit(struct brw_context *brw,
if (unlikely(cfg == NULL)) {
assert(fail_msg);
prog->LinkStatus = false;
ralloc_strcat(&prog->InfoLog, fail_msg);
_mesa_problem(NULL, "Failed to compile compute shader: %s\n",
fail_msg);
if (error_str)
*error_str = ralloc_strdup(mem_ctx, fail_msg);
return NULL;
}
fs_generator g(brw->intelScreen->compiler, brw,
mem_ctx, (void*) key, &prog_data->base,
fs_generator g(compiler, log_data, mem_ctx, (void*) key, &prog_data->base,
v8.promoted_constants, v8.runtime_check_aads_emit, "CS");
if (INTEL_DEBUG & DEBUG_CS) {
char *name = ralloc_asprintf(mem_ctx, "%s compute shader %d",
prog->Label ? prog->Label : "unnamed",
prog->Name);
char *name = ralloc_asprintf(mem_ctx, "%s compute shader %s",
shader->info.label ? shader->info.label :
"unnamed",
shader->info.name);
g.enable_debug(name);
}