fdo-mirrors/mesa
1
0
Fork 0
mirror of https://gitlab.freedesktop.org/mesa/mesa.git synced 2026-05-08 04:48:08 +02:00
Code Issues Projects Releases Packages Wiki Activity Actions 1

radeonsi: use default float mode for CL

Browse source 
Signed-off-by: Karol Herbst <kherbst@redhat.com>
Reviewed-by: Pierre-Eric Pelloux-Prayer <pierre-eric.pelloux-prayer@amd.com>
Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/19232>
This commit is contained in:
Karol Herbst 2022-09-29 03:00:38 +02:00 • committed by Marge Bot
parent e5ef95e31f
commit c9ab7d7525
3 changed files with 10 additions and 5 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

3
src/gallium/drivers/radeonsi/si_shader.c
View file

@ -2461,7 +2461,8 @@ si_get_shader_part(struct si_screen *sscreen, struct si_shader_part **list,
}
struct si_shader_context ctx;
si_llvm_context_init(&ctx, sscreen, compiler, wave32 ? 32 : 64, exports_color_null, exports_mrtz);
si_llvm_context_init(&ctx, sscreen, compiler, wave32 ? 32 : 64, exports_color_null, exports_mrtz,
AC_FLOAT_MODE_DEFAULT_OPENGL);
ctx.shader = &shader;
ctx.stage = stage;

3
src/gallium/drivers/radeonsi/si_shader_internal.h
View file

@ -182,7 +182,8 @@ bool si_compile_llvm(struct si_screen *sscreen, struct si_shader_binary *binary,
gl_shader_stage stage, const char *name, bool less_optimized);
void si_llvm_context_init(struct si_shader_context *ctx, struct si_screen *sscreen,
struct ac_llvm_compiler *compiler, unsigned wave_size,
bool exports_color_null, bool exports_mrtz);
bool exports_color_null, bool exports_mrtz,
enum ac_float_mode float_mode);
void si_llvm_create_func(struct si_shader_context *ctx, const char *name, LLVMTypeRef *return_types,
unsigned num_return_elems, unsigned max_workgroup_size);
void si_llvm_create_main_func(struct si_shader_context *ctx);

9
src/gallium/drivers/radeonsi/si_shader_llvm.c
View file

@ -126,14 +126,15 @@ bool si_compile_llvm(struct si_screen *sscreen, struct si_shader_binary *binary,
void si_llvm_context_init(struct si_shader_context *ctx, struct si_screen *sscreen,
struct ac_llvm_compiler *compiler, unsigned wave_size,
bool exports_color_null, bool exports_mrtz)
bool exports_color_null, bool exports_mrtz,
enum ac_float_mode float_mode)
{
memset(ctx, 0, sizeof(*ctx));
ctx->screen = sscreen;
ctx->compiler = compiler;
ac_llvm_context_init(&ctx->ac, compiler, sscreen->info.gfx_level, sscreen->info.family,
sscreen->info.has_3d_cube_border_color_mipmap, AC_FLOAT_MODE_DEFAULT_OPENGL,
sscreen->info.has_3d_cube_border_color_mipmap, float_mode,
wave_size, 64, exports_color_null, exports_mrtz);
}
@ -1091,13 +1092,15 @@ bool si_llvm_compile_shader(struct si_screen *sscreen, struct ac_llvm_compiler *
{
struct si_shader_selector *sel = shader->selector;
struct si_shader_context ctx;
enum ac_float_mode float_mode = nir->info.stage == MESA_SHADER_KERNEL ? AC_FLOAT_MODE_DEFAULT : AC_FLOAT_MODE_DEFAULT_OPENGL;
bool exports_color_null = sel->info.colors_written;
bool exports_mrtz = sel->info.writes_z || sel->info.writes_stencil || sel->info.writes_samplemask;
if (!exports_mrtz && !exports_color_null)
exports_color_null = si_shader_uses_discard(shader) || sscreen->info.gfx_level < GFX10;
si_llvm_context_init(&ctx, sscreen, compiler, shader->wave_size, exports_color_null, exports_mrtz);
si_llvm_context_init(&ctx, sscreen, compiler, shader->wave_size, exports_color_null, exports_mrtz,
float_mode);
ctx.args = args;
if (!si_llvm_translate_nir(&ctx, shader, nir, false)) {