X-Git-Url: https://git.libre-soc.org/?a=blobdiff_plain;f=src%2Famd%2Fvulkan%2Fradv_shader.h;h=59497891ffd219912317fea5a3fec8a33e5c1c99;hb=HEAD;hp=66cd005e5fc45f4dccba196fcd0c8ba71b265816;hpb=43f2f01cc89def665bd0e33f9ad689825b85e977;p=mesa.git diff --git a/src/amd/vulkan/radv_shader.h b/src/amd/vulkan/radv_shader.h index 66cd005e5fc..59497891ffd 100644 --- a/src/amd/vulkan/radv_shader.h +++ b/src/amd/vulkan/radv_shader.h @@ -28,27 +28,20 @@ #ifndef RADV_SHADER_H #define RADV_SHADER_H -#include "radv_debug.h" -#include "radv_private.h" +#include "ac_binary.h" +#include "amd_family.h" +#include "radv_constants.h" #include "nir/nir.h" +#include "vulkan/vulkan.h" +#include "vulkan/util/vk_object.h" -/* descriptor index into scratch ring offsets */ -#define RING_SCRATCH 0 -#define RING_ESGS_VS 1 -#define RING_ESGS_GS 2 -#define RING_GSVS_VS 3 -#define RING_GSVS_GS 4 -#define RING_HS_TESS_FACTOR 5 -#define RING_HS_TESS_OFFCHIP 6 -#define RING_PS_SAMPLE_POSITIONS 7 +#define RADV_VERT_ATTRIB_MAX MAX2(VERT_ATTRIB_MAX, VERT_ATTRIB_GENERIC0 + MAX_VERTEX_ATTRIBS) -// Match MAX_SETS from radv_descriptor_set.h -#define RADV_UD_MAX_SETS MAX_SETS - -#define RADV_NUM_PHYSICAL_VGPRS 256 +struct radv_device; struct radv_shader_module { + struct vk_object_base base; struct nir_shader *nir; unsigned char sha1[20]; uint32_t size; @@ -62,7 +55,20 @@ enum { RADV_ALPHA_ADJUST_SSCALED = 3, }; +struct radv_vs_out_key { + uint32_t as_es:1; + uint32_t as_ls:1; + uint32_t as_ngg:1; + uint32_t as_ngg_passthrough:1; + uint32_t export_prim_id:1; + uint32_t export_layer_id:1; + uint32_t export_clip_dists:1; + uint32_t export_viewport_index:1; +}; + struct radv_vs_variant_key { + struct radv_vs_out_key out; + uint32_t instance_rate_inputs; uint32_t instance_rate_divisors[MAX_VERTEX_ATTRIBS]; uint8_t vertex_attribute_formats[MAX_VERTEX_ATTRIBS]; @@ -77,18 +83,13 @@ struct radv_vs_variant_key { /* For some formats the channels have to be shuffled. */ uint32_t post_shuffle; - uint32_t as_es:1; - uint32_t as_ls:1; - uint32_t export_prim_id:1; - uint32_t export_layer_id:1; - uint32_t export_clip_dists:1; + /* Output primitive type. */ + uint8_t outprim; }; struct radv_tes_variant_key { - uint32_t as_es:1; - uint32_t export_prim_id:1; - uint32_t export_layer_id:1; - uint32_t export_clip_dists:1; + struct radv_vs_out_key out; + uint8_t num_patches; uint8_t tcs_num_outputs; }; @@ -107,6 +108,11 @@ struct radv_fs_variant_key { uint8_t num_samples; uint32_t is_int8; uint32_t is_int10; + bool is_dual_src; +}; + +struct radv_cs_variant_key { + uint8_t subgroup_size; }; struct radv_shader_variant_key { @@ -115,24 +121,45 @@ struct radv_shader_variant_key { struct radv_fs_variant_key fs; struct radv_tes_variant_key tes; struct radv_tcs_variant_key tcs; + struct radv_cs_variant_key cs; + + /* A common prefix of the vs and tes keys. */ + struct radv_vs_out_key vs_common_out; }; bool has_multiview_view_index; }; +enum radv_compiler_debug_level { + RADV_COMPILER_DEBUG_LEVEL_PERFWARN, + RADV_COMPILER_DEBUG_LEVEL_ERROR, +}; + struct radv_nir_compiler_options { struct radv_pipeline_layout *layout; struct radv_shader_variant_key key; - bool unsafe_math; - bool supports_spill; + bool explicit_scratch_args; bool clamp_shadow_reference; + bool robust_buffer_access; bool dump_shader; bool dump_preoptir; - bool record_llvm_ir; + bool record_ir; + bool record_stats; bool check_ir; + bool has_ls_vgpr_init_bug; + bool use_ngg_streamout; + bool enable_mrt_output_nan_fixup; + bool disable_optimizations; /* only used by ACO */ enum radeon_family family; enum chip_class chip_class; uint32_t tess_offchip_block_dw_size; uint32_t address32_hi; + + struct { + void (*func)(void *private_data, + enum radv_compiler_debug_level level, + const char *message); + void *private_data; + } debug; }; enum radv_ud_index { @@ -142,7 +169,8 @@ enum radv_ud_index { AC_UD_INDIRECT_DESCRIPTOR_SETS = 3, AC_UD_VIEW_INDEX = 4, AC_UD_STREAMOUT_BUFFERS = 5, - AC_UD_SHADER_START = 6, + AC_UD_NGG_GS_STATE = 6, + AC_UD_SHADER_START = 7, AC_UD_VS_VERTEX_BUFFERS = AC_UD_SHADER_START, AC_UD_VS_BASE_VERTEX_START_INSTANCE, AC_UD_VS_MAX_UD, @@ -170,6 +198,51 @@ struct radv_streamout_info { uint32_t enabled_stream_buffers_mask; }; +struct radv_userdata_info { + int8_t sgpr_idx; + uint8_t num_sgprs; +}; + +struct radv_userdata_locations { + struct radv_userdata_info descriptor_sets[MAX_SETS]; + struct radv_userdata_info shader_data[AC_UD_MAX_UD]; + uint32_t descriptor_sets_enabled; +}; + +struct radv_vs_output_info { + uint8_t vs_output_param_offset[VARYING_SLOT_MAX]; + uint8_t clip_dist_mask; + uint8_t cull_dist_mask; + uint8_t param_exports; + bool writes_pointsize; + bool writes_layer; + bool writes_viewport_index; + bool export_prim_id; + unsigned pos_exports; +}; + +struct radv_es_output_info { + uint32_t esgs_itemsize; +}; + +struct gfx9_gs_info { + uint32_t vgt_gs_onchip_cntl; + uint32_t vgt_gs_max_prims_per_subgroup; + uint32_t vgt_esgs_ring_itemsize; + uint32_t lds_size; +}; + +struct gfx10_ngg_info { + uint16_t ngg_emit_size; /* in dwords */ + uint32_t hw_max_esverts; + uint32_t max_gsprims; + uint32_t max_out_verts; + uint32_t prim_amp_factor; + uint32_t vgt_esgs_ring_itemsize; + uint32_t esgs_ring_size; + bool max_vert_out_per_gs_instance; +}; + struct radv_shader_info { bool loads_push_constants; bool loads_dynamic_offsets; @@ -183,27 +256,62 @@ struct radv_shader_info { bool needs_multiview_view_index; bool uses_invocation_id; bool uses_prim_id; + uint8_t wave_size; + uint8_t ballot_bit_size; + struct radv_userdata_locations user_sgprs_locs; + unsigned num_user_sgprs; + unsigned num_input_sgprs; + unsigned num_input_vgprs; + unsigned private_mem_vgprs; + bool need_indirect_descriptor_sets; + bool is_ngg; + bool is_ngg_passthrough; struct { uint64_t ls_outputs_written; - uint8_t input_usage_mask[VERT_ATTRIB_MAX]; + uint8_t input_usage_mask[RADV_VERT_ATTRIB_MAX]; uint8_t output_usage_mask[VARYING_SLOT_VAR31 + 1]; bool has_vertex_buffers; /* needs vertex buffers and base/start */ bool needs_draw_id; bool needs_instance_id; + struct radv_vs_output_info outinfo; + struct radv_es_output_info es_info; + bool as_es; + bool as_ls; + bool export_prim_id; + uint8_t num_linked_outputs; } vs; struct { uint8_t output_usage_mask[VARYING_SLOT_VAR31 + 1]; uint8_t num_stream_output_components[4]; uint8_t output_streams[VARYING_SLOT_VAR31 + 1]; uint8_t max_stream; + bool writes_memory; + unsigned gsvs_vertex_size; + unsigned max_gsvs_emit_size; + unsigned vertices_in; + unsigned vertices_out; + unsigned output_prim; + unsigned invocations; + unsigned es_type; /* GFX9: VS or TES */ + uint8_t num_linked_inputs; } gs; struct { uint8_t output_usage_mask[VARYING_SLOT_VAR31 + 1]; + struct radv_vs_output_info outinfo; + struct radv_es_output_info es_info; + bool as_es; + unsigned primitive_mode; + enum gl_tess_spacing spacing; + bool ccw; + bool point_mode; + bool export_prim_id; + uint8_t num_linked_inputs; + uint8_t num_linked_patch_inputs; + uint8_t num_linked_outputs; } tes; struct { bool force_persample; bool needs_sample_positions; - bool uses_input_attachments; bool writes_memory; bool writes_z; bool writes_stencil; @@ -211,101 +319,93 @@ struct radv_shader_info { bool has_pcoord; bool prim_id_input; bool layer_input; + bool viewport_index_input; uint8_t num_input_clips_culls; + uint32_t input_mask; + uint32_t flat_shaded_mask; + uint32_t explicit_shaded_mask; + uint32_t float16_shaded_mask; + uint32_t num_interp; + uint32_t cb_shader_mask; + bool can_discard; + bool early_fragment_test; + bool post_depth_coverage; + uint8_t depth_layout; } ps; struct { bool uses_grid_size; bool uses_block_id[3]; bool uses_thread_id[3]; bool uses_local_invocation_idx; + unsigned block_size[3]; } cs; struct { uint64_t outputs_written; uint64_t patch_outputs_written; + uint64_t tes_inputs_read; + uint64_t tes_patch_inputs_read; + unsigned tcs_vertices_out; + uint32_t num_patches; + uint32_t num_lds_blocks; + uint8_t num_linked_inputs; + uint8_t num_linked_outputs; + uint8_t num_linked_patch_outputs; } tcs; struct radv_streamout_info so; + + struct gfx9_gs_info gs_ring_info; + struct gfx10_ngg_info ngg_info; + + unsigned float_controls_mode; }; -struct radv_userdata_info { - int8_t sgpr_idx; - uint8_t num_sgprs; +enum radv_shader_binary_type { + RADV_BINARY_TYPE_LEGACY, + RADV_BINARY_TYPE_RTLD }; -struct radv_userdata_locations { - struct radv_userdata_info descriptor_sets[RADV_UD_MAX_SETS]; - struct radv_userdata_info shader_data[AC_UD_MAX_UD]; - uint32_t descriptor_sets_enabled; +struct radv_shader_binary { + enum radv_shader_binary_type type; + gl_shader_stage stage; + bool is_gs_copy_shader; + + struct radv_shader_info info; + + /* Self-referential size so we avoid consistency issues. */ + uint32_t total_size; }; -struct radv_vs_output_info { - uint8_t vs_output_param_offset[VARYING_SLOT_MAX]; - uint8_t clip_dist_mask; - uint8_t cull_dist_mask; - uint8_t param_exports; - bool writes_pointsize; - bool writes_layer; - bool writes_viewport_index; - bool export_prim_id; - unsigned pos_exports; +struct radv_shader_binary_legacy { + struct radv_shader_binary base; + struct ac_shader_config config; + unsigned code_size; + unsigned exec_size; + unsigned ir_size; + unsigned disasm_size; + unsigned stats_size; + + /* data has size of stats_size + code_size + ir_size + disasm_size + 2, + * where the +2 is for 0 of the ir strings. */ + uint8_t data[0]; }; -struct radv_es_output_info { - uint32_t esgs_itemsize; +struct radv_shader_binary_rtld { + struct radv_shader_binary base; + unsigned elf_size; + unsigned llvm_ir_size; + uint8_t data[0]; }; -struct radv_shader_variant_info { - struct radv_userdata_locations user_sgprs_locs; - struct radv_shader_info info; - unsigned num_user_sgprs; - unsigned num_input_sgprs; - unsigned num_input_vgprs; - unsigned private_mem_vgprs; - bool need_indirect_descriptor_sets; - struct { - struct { - struct radv_vs_output_info outinfo; - struct radv_es_output_info es_info; - bool as_es; - bool as_ls; - bool export_prim_id; - } vs; - struct { - unsigned num_interp; - uint32_t input_mask; - uint32_t flat_shaded_mask; - uint32_t float16_shaded_mask; - bool can_discard; - bool early_fragment_test; - } fs; - struct { - unsigned block_size[3]; - } cs; - struct { - unsigned vertices_in; - unsigned vertices_out; - unsigned output_prim; - unsigned invocations; - unsigned gsvs_vertex_size; - unsigned max_gsvs_emit_size; - unsigned es_type; /* GFX9: VS or TES */ - } gs; - struct { - unsigned tcs_vertices_out; - uint32_t num_patches; - uint32_t lds_size; - } tcs; - struct { - struct radv_vs_output_info outinfo; - struct radv_es_output_info es_info; - bool as_es; - unsigned primitive_mode; - enum gl_tess_spacing spacing; - bool ccw; - bool point_mode; - bool export_prim_id; - } tes; - }; +struct radv_compiler_statistic_info { + char name[32]; + char desc[64]; +}; + +struct radv_compiler_statistics { + unsigned count; + struct radv_compiler_statistic_info *infos; + uint32_t values[]; }; struct radv_shader_variant { @@ -315,14 +415,16 @@ struct radv_shader_variant { uint64_t bo_offset; struct ac_shader_config config; uint32_t code_size; - struct radv_shader_variant_info info; + uint32_t exec_size; + struct radv_shader_info info; /* debug only */ - uint32_t *spirv; + char *spirv; uint32_t spirv_size; - struct nir_shader *nir; + char *nir_string; char *disasm_string; - char *llvm_ir_string; + char *ir_string; + struct radv_compiler_statistics *statistics; struct list_head slab_list; }; @@ -349,65 +451,84 @@ radv_shader_compile_to_nir(struct radv_device *device, gl_shader_stage stage, const VkSpecializationInfo *spec_info, const VkPipelineCreateFlags flags, - const struct radv_pipeline_layout *layout); - -void * -radv_alloc_shader_memory(struct radv_device *device, - struct radv_shader_variant *shader); + const struct radv_pipeline_layout *layout, + unsigned subgroup_size, unsigned ballot_bit_size); void radv_destroy_shader_slabs(struct radv_device *device); +VkResult +radv_create_shaders(struct radv_pipeline *pipeline, + struct radv_device *device, + struct radv_pipeline_cache *cache, + const struct radv_pipeline_key *key, + const VkPipelineShaderStageCreateInfo **pStages, + const VkPipelineCreateFlags flags, + VkPipelineCreationFeedbackEXT *pipeline_feedback, + VkPipelineCreationFeedbackEXT **stage_feedbacks); + struct radv_shader_variant * radv_shader_variant_create(struct radv_device *device, - struct radv_shader_module *module, - struct nir_shader *const *shaders, - int shader_count, - struct radv_pipeline_layout *layout, - const struct radv_shader_variant_key *key, - void **code_out, - unsigned *code_size_out); + const struct radv_shader_binary *binary, + bool keep_shader_info); +struct radv_shader_variant * +radv_shader_variant_compile(struct radv_device *device, + struct radv_shader_module *module, + struct nir_shader *const *shaders, + int shader_count, + struct radv_pipeline_layout *layout, + const struct radv_shader_variant_key *key, + struct radv_shader_info *info, + bool keep_shader_info, bool keep_statistic_info, + bool disable_optimizations, + struct radv_shader_binary **binary_out); struct radv_shader_variant * radv_create_gs_copy_shader(struct radv_device *device, struct nir_shader *nir, - void **code_out, unsigned *code_size_out, - bool multiview); + struct radv_shader_info *info, + struct radv_shader_binary **binary_out, + bool multiview, bool keep_shader_info, + bool keep_statistic_info, + bool disable_optimizations); + +struct radv_shader_variant * +radv_create_trap_handler_shader(struct radv_device *device); void radv_shader_variant_destroy(struct radv_device *device, struct radv_shader_variant *variant); -const char * -radv_get_shader_name(struct radv_shader_variant *var, gl_shader_stage stage); -void -radv_shader_dump_stats(struct radv_device *device, - struct radv_shader_variant *variant, - gl_shader_stage stage, - FILE *file); +unsigned +radv_get_max_waves(struct radv_device *device, + struct radv_shader_variant *variant, + gl_shader_stage stage); + +unsigned +radv_get_max_workgroup_size(enum chip_class chip_class, + gl_shader_stage stage, + const unsigned *sizes); -static inline bool +const char * +radv_get_shader_name(struct radv_shader_info *info, + gl_shader_stage stage); + +bool radv_can_dump_shader(struct radv_device *device, struct radv_shader_module *module, - bool is_gs_copy_shader) -{ - if (!(device->instance->debug_flags & RADV_DEBUG_DUMP_SHADERS)) - return false; + bool is_gs_copy_shader); - /* Only dump non-meta shaders, useful for debugging purposes. */ - return (module && !module->nir) || is_gs_copy_shader; -} - -static inline bool +bool radv_can_dump_shader_stats(struct radv_device *device, - struct radv_shader_module *module) -{ - /* Only dump non-meta shader stats. */ - return device->instance->debug_flags & RADV_DEBUG_DUMP_SHADER_STATS && - module && !module->nir; -} + struct radv_shader_module *module); + +VkResult +radv_dump_shader_stats(struct radv_device *device, + struct radv_pipeline *pipeline, + gl_shader_stage stage, FILE *output); -static inline unsigned shader_io_get_unique_index(gl_varying_slot slot) +static inline unsigned +shader_io_get_unique_index(gl_varying_slot slot) { /* handle patch indices separate */ if (slot == VARYING_SLOT_TESS_LEVEL_OUTER) @@ -430,4 +551,89 @@ static inline unsigned shader_io_get_unique_index(gl_varying_slot slot) unreachable("illegal slot in get unique index\n"); } +static inline unsigned +calculate_tess_lds_size(enum chip_class chip_class, + unsigned tcs_num_input_vertices, + unsigned tcs_num_output_vertices, + unsigned tcs_num_inputs, + unsigned tcs_num_patches, + unsigned tcs_num_outputs, + unsigned tcs_num_patch_outputs) +{ + unsigned input_vertex_size = tcs_num_inputs * 16; + unsigned output_vertex_size = tcs_num_outputs * 16; + + unsigned input_patch_size = tcs_num_input_vertices * input_vertex_size; + + unsigned pervertex_output_patch_size = tcs_num_output_vertices * output_vertex_size; + unsigned output_patch_size = pervertex_output_patch_size + tcs_num_patch_outputs * 16; + + unsigned output_patch0_offset = input_patch_size * tcs_num_patches; + + unsigned lds_size = output_patch0_offset + output_patch_size * tcs_num_patches; + + if (chip_class >= GFX7) { + assert(lds_size <= 65536); + lds_size = align(lds_size, 512) / 512; + } else { + assert(lds_size <= 32768); + lds_size = align(lds_size, 256) / 256; + } + + return lds_size; +} + +static inline unsigned +get_tcs_num_patches(unsigned tcs_num_input_vertices, + unsigned tcs_num_output_vertices, + unsigned tcs_num_inputs, + unsigned tcs_num_outputs, + unsigned tcs_num_patch_outputs, + unsigned tess_offchip_block_dw_size, + enum chip_class chip_class, + enum radeon_family family) +{ + uint32_t input_vertex_size = tcs_num_inputs * 16; + uint32_t input_patch_size = tcs_num_input_vertices * input_vertex_size; + uint32_t output_vertex_size = tcs_num_outputs * 16; + uint32_t pervertex_output_patch_size = tcs_num_output_vertices * output_vertex_size; + uint32_t output_patch_size = pervertex_output_patch_size + tcs_num_patch_outputs * 16; + + /* Ensure that we only need one wave per SIMD so we don't need to check + * resource usage. Also ensures that the number of tcs in and out + * vertices per threadgroup are at most 256. + */ + unsigned num_patches = 64 / MAX2(tcs_num_input_vertices, tcs_num_output_vertices) * 4; + /* Make sure that the data fits in LDS. This assumes the shaders only + * use LDS for the inputs and outputs. + */ + unsigned hardware_lds_size = 32768; + + /* Looks like STONEY hangs if we use more than 32 KiB LDS in a single + * threadgroup, even though there is more than 32 KiB LDS. + * + * Test: dEQP-VK.tessellation.shader_input_output.barrier + */ + if (chip_class >= GFX7 && family != CHIP_STONEY) + hardware_lds_size = 65536; + + num_patches = MIN2(num_patches, hardware_lds_size / (input_patch_size + output_patch_size)); + /* Make sure the output data fits in the offchip buffer */ + num_patches = MIN2(num_patches, (tess_offchip_block_dw_size * 4) / output_patch_size); + /* Not necessary for correctness, but improves performance. The + * specific value is taken from the proprietary driver. + */ + num_patches = MIN2(num_patches, 40); + + /* GFX6 bug workaround - limit LS-HS threadgroups to only one wave. */ + if (chip_class == GFX6) { + unsigned one_wave = 64 / MAX2(tcs_num_input_vertices, tcs_num_output_vertices); + num_patches = MIN2(num_patches, one_wave); + } + return num_patches; +} + +void +radv_lower_fs_io(nir_shader *nir); + #endif