radv: emit framebuffer state from primary if secondary doesn't inherit it

[mesa.git] / src / amd / vulkan / radv_nir_to_llvm.c

diff --git a/src/amd/vulkan/radv_nir_to_llvm.c b/src/amd/vulkan/radv_nir_to_llvm.c
index dca4bebcdd16cc87e854705617c691744ad78633..d6f286fe4ec6af16cfbe022f25045101838816d1 100644 (file)
--- a/src/amd/vulkan/radv_nir_to_llvm.c
+++ b/src/amd/vulkan/radv_nir_to_llvm.c
@@ -36,7 +36,6 @@
 #include <llvm-c/Transforms/Utils.h>
 
 #include "sid.h"
-#include "gfx9d.h"
 #include "ac_binary.h"
 #include "ac_llvm_util.h"
 #include "ac_llvm_build.h"
@@ -518,11 +517,8 @@ create_llvm_function(LLVMContextRef ctx, LLVMModuleRef module,
                                                     options->address32_hi);
        }
 
-       if (max_workgroup_size) {
-               ac_llvm_add_target_dep_function_attr(main_function,
-                                                    "amdgpu-max-work-group-size",
-                                                    max_workgroup_size);
-       }
+       ac_llvm_set_workgroup_size(main_function, max_workgroup_size);
+
        if (options->unsafe_math) {
                /* These were copied from some LLVM test. */
                LLVMAddTargetDependentFunctionAttr(main_function,
@@ -1589,13 +1585,13 @@ store_tcs_output(struct ac_shader_abi *abi,
                if (!is_tess_factor && writemask != 0xF)
                        ac_build_buffer_store_dword(&ctx->ac, ctx->hs_ring_tess_offchip, value, 1,
                                                    buf_addr, ctx->oc_lds,
-                                                   4 * (base + chan), 1, 0, true, false);
+                                                   4 * (base + chan), 1, 0, false);
        }
 
        if (writemask == 0xF) {
                ac_build_buffer_store_dword(&ctx->ac, ctx->hs_ring_tess_offchip, src, 4,
                                            buf_addr, ctx->oc_lds,
-                                           (base * 4), 1, 0, true, false);
+                                           (base * 4), 1, 0, false);
        }
 }
 
@@ -1862,7 +1858,7 @@ visit_emit_vertex(struct ac_shader_abi *abi, unsigned stream, LLVMValueRef *addr
                                                    ctx->gsvs_ring[stream],
                                                    out_val, 1,
                                                    voffset, ctx->gs2vs_offset, 0,
-                                                   1, 1, true, true);
+                                                   1, 1, true);
                }
        }
 
@@ -2768,7 +2764,7 @@ radv_emit_stream_output(struct radv_shader_context *ctx,
                /* fall through */
        case 4: /* as v4i32 */
                vdata = ac_build_gather_values(&ctx->ac, out,
-                                              HAVE_LLVM < 0x900 ?
+                                              !ac_has_vec3_support(ctx->ac.chip_class, false) ?
                                               util_next_power_of_two(num_comps) :
                                               num_comps);
                break;
@@ -2777,7 +2773,7 @@ radv_emit_stream_output(struct radv_shader_context *ctx,
        ac_build_buffer_store_dword(&ctx->ac, so_buffers[buf],
                                    vdata, num_comps, so_write_offsets[buf],
                                    ctx->ac.i32_0, offset,
-                                   1, 1, true, false);
+                                   1, 1, false);
 }
 
 static void
@@ -3154,7 +3150,7 @@ handle_es_outputs_post(struct radv_shader_context *ctx,
                                                            out_val, 1,
                                                            NULL, ctx->es2gs_offset,
                                                            (4 * param_index + j) * 4,
-                                                           1, 1, true, true);
+                                                           1, 1, true);
                        }
                }
        }
@@ -3287,7 +3283,7 @@ write_tess_factors(struct radv_shader_context *ctx)
                ac_build_buffer_store_dword(&ctx->ac, buffer,
                                            LLVMConstInt(ctx->ac.i32, 0x80000000, false),
                                            1, ctx->ac.i32_0, tf_base,
-                                           0, 1, 0, true, false);
+                                           0, 1, 0, false);
                tf_offset += 4;
 
                ac_nir_build_endif(&inner_if_ctx);
@@ -3296,11 +3292,11 @@ write_tess_factors(struct radv_shader_context *ctx)
        /* Store the tessellation factors. */
        ac_build_buffer_store_dword(&ctx->ac, buffer, vec0,
                                    MIN2(stride, 4), byteoffset, tf_base,
-                                   tf_offset, 1, 0, true, false);
+                                   tf_offset, 1, 0, false);
        if (vec1)
                ac_build_buffer_store_dword(&ctx->ac, buffer, vec1,
                                            stride - 4, byteoffset, tf_base,
-                                           16 + tf_offset, 1, 0, true, false);
+                                           16 + tf_offset, 1, 0, false);
 
        //store to offchip for TES to read - only if TES reads them
        if (ctx->options->key.tcs.tes_reads_tess_factors) {
@@ -3317,7 +3313,7 @@ write_tess_factors(struct radv_shader_context *ctx)
 
                ac_build_buffer_store_dword(&ctx->ac, ctx->hs_ring_tess_offchip, outer_vec,
                                            outer_comps, tf_outer_offset,
-                                           ctx->oc_lds, 0, 1, 0, true, false);
+                                           ctx->oc_lds, 0, 1, 0, false);
                if (inner_comps) {
                        param_inner = shader_io_get_unique_index(VARYING_SLOT_TESS_LEVEL_INNER);
                        tf_inner_offset = get_tcs_tes_buffer_address(ctx, NULL,
@@ -3327,7 +3323,7 @@ write_tess_factors(struct radv_shader_context *ctx)
                                ac_build_gather_values(&ctx->ac, inner, inner_comps);
                        ac_build_buffer_store_dword(&ctx->ac, ctx->hs_ring_tess_offchip, inner_vec,
                                                    inner_comps, tf_inner_offset,
-                                                   ctx->oc_lds, 0, 1, 0, true, false);
+                                                   ctx->oc_lds, 0, 1, 0, false);
                }
        }
        ac_nir_build_endif(&if_ctx);