Mesa (master): ac/nir: Add LS-HS input VGPR workaround.

[Bas Nieuwenhuizen]

Module: Mesa
Branch: master
Commit: 640f2c458f96c540a703740312da0fd21c28a908
URL:    http://cgit.freedesktop.org/mesa/mesa/commit/?id=640f2c458f96c540a703740312da0fd21c28a908

Author: Bas Nieuwenhuizen <bas at basnieuwenhuizen.nl>
Date:   Thu Oct 19 02:58:34 2017 +0200

ac/nir: Add LS-HS input VGPR workaround.

Reviewed-by: Dave Airlie <airlied at redhat.com>

---

 src/amd/common/ac_nir_to_llvm.c | 18 ++++++++++++++++++
 1 file changed, 18 insertions(+)

diff --git a/src/amd/common/ac_nir_to_llvm.c b/src/amd/common/ac_nir_to_llvm.c
index cb011bd88b..242675654d 100644
--- a/src/amd/common/ac_nir_to_llvm.c
+++ b/src/amd/common/ac_nir_to_llvm.c
@@ -6377,6 +6377,20 @@ ac_nir_get_max_workgroup_size(enum chip_class chip_class,
 	return max_workgroup_size;
 }
 
+/* Fixup the HW not emitting the TCS regs if there are no HS threads. */
+static void ac_nir_fixup_ls_hs_input_vgprs(struct nir_to_llvm_context *ctx)
+{
+	LLVMValueRef count = ac_build_bfe(&ctx->ac, ctx->merged_wave_info,
+	                                  LLVMConstInt(ctx->ac.i32, 8, false),
+	                                  LLVMConstInt(ctx->ac.i32, 8, false), false);
+	LLVMValueRef hs_empty = LLVMBuildICmp(ctx->ac.builder, LLVMIntEQ, count,
+	                                      LLVMConstInt(ctx->ac.i32, 0, false), "");
+	ctx->abi.instance_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->rel_auto_id, ctx->abi.instance_id, "");
+	ctx->vs_prim_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->abi.vertex_id, ctx->vs_prim_id, "");
+	ctx->rel_auto_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->tcs_rel_ids, ctx->rel_auto_id, "");
+	ctx->abi.vertex_id = LLVMBuildSelect(ctx->ac.builder, hs_empty, ctx->tcs_patch_id, ctx->abi.vertex_id, "");
+}
+
 void ac_nir_translate(struct ac_llvm_context *ac, struct ac_shader_abi *abi,
 		      struct nir_shader *nir, struct nir_to_llvm_context *nctx)
 {
@@ -6474,6 +6488,10 @@ LLVMModuleRef ac_translate_nir_to_llvm(LLVMTargetMachineRef tm,
 	ctx.abi.load_ssbo = radv_load_ssbo;
 	ctx.abi.load_sampler_desc = radv_get_sampler_desc;
 
+	if (ctx.ac.chip_class == GFX9 &&
+	    shaders[shader_count - 1]->stage == MESA_SHADER_TESS_CTRL)
+		ac_nir_fixup_ls_hs_input_vgprs(&ctx);
+
 	for(int i = 0; i < shader_count; ++i) {
 		ctx.stage = shaders[i]->stage;
 		ctx.output_mask = 0;