1/* 2 * Copyright 2018 Collabora Ltd. 3 * 4 * Permission is hereby granted, free of charge, to any person obtaining a 5 * copy of this software and associated documentation files (the "Software"), 6 * to deal in the Software without restriction, including without limitation 7 * on the rights to use, copy, modify, merge, publish, distribute, sub 8 * license, and/or sell copies of the Software, and to permit persons to whom 9 * the Software is furnished to do so, subject to the following conditions: 10 * 11 * The above copyright notice and this permission notice (including the next 12 * paragraph) shall be included in all copies or substantial portions of the 13 * Software. 14 * 15 * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR 16 * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, 17 * FITNESS FOR A PARTICULAR PURPOSE AND NON-INFRINGEMENT. IN NO EVENT SHALL 18 * THE AUTHOR(S) AND/OR THEIR SUPPLIERS BE LIABLE FOR ANY CLAIM, 19 * DAMAGES OR OTHER LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR 20 * OTHERWISE, ARISING FROM, OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE 21 * USE OR OTHER DEALINGS IN THE SOFTWARE. 22 */ 23 24#include "zink_program.h" 25 26#include "zink_compiler.h" 27#include "zink_context.h" 28#include "zink_descriptors.h" 29#include "zink_helpers.h" 30#include "zink_render_pass.h" 31#include "zink_resource.h" 32#include "zink_screen.h" 33#include "zink_state.h" 34#include "zink_inlines.h" 35 36#include "util/hash_table.h" 37#include "util/set.h" 38#include "util/u_debug.h" 39#include "util/u_memory.h" 40#include "util/u_prim.h" 41#include "tgsi/tgsi_from_mesa.h" 42 43/* for pipeline cache */ 44#define XXH_INLINE_ALL 45#include "util/xxhash.h" 46 47struct gfx_pipeline_cache_entry { 48 struct zink_gfx_pipeline_state state; 49 VkPipeline pipeline; 50}; 51 52struct compute_pipeline_cache_entry { 53 struct zink_compute_pipeline_state state; 54 VkPipeline pipeline; 55}; 56 57void 58debug_describe_zink_gfx_program(char *buf, const struct zink_gfx_program *ptr) 59{ 60 sprintf(buf, "zink_gfx_program"); 61} 62 63void 64debug_describe_zink_compute_program(char *buf, const struct zink_compute_program *ptr) 65{ 66 sprintf(buf, "zink_compute_program"); 67} 68 69static bool 70shader_key_matches(const struct zink_shader_module *zm, const struct zink_shader_key *key, unsigned num_uniforms) 71{ 72 if (zm->key_size != key->size || zm->num_uniforms != num_uniforms) 73 return false; 74 return !memcmp(zm->key, key, zm->key_size) && 75 (!num_uniforms || !memcmp(zm->key + zm->key_size, key->base.inlined_uniform_values, zm->num_uniforms * sizeof(uint32_t))); 76} 77 78static uint32_t 79shader_module_hash(const struct zink_shader_module *zm) 80{ 81 unsigned key_size = zm->key_size + zm->num_uniforms * sizeof(uint32_t); 82 return _mesa_hash_data(zm->key, key_size); 83} 84 85static struct zink_shader_module * 86get_shader_module_for_stage(struct zink_context *ctx, struct zink_screen *screen, 87 struct zink_shader *zs, struct zink_gfx_program *prog, 88 struct zink_gfx_pipeline_state *state) 89{ 90 gl_shader_stage stage = zs->nir->info.stage; 91 enum pipe_shader_type pstage = pipe_shader_type_from_mesa(stage); 92 VkShaderModule mod; 93 struct zink_shader_module *zm = NULL; 94 unsigned base_size = 0; 95 struct zink_shader_key *key = &state->shader_keys.key[pstage]; 96 97 if (ctx && zs->nir->info.num_inlinable_uniforms && 98 ctx->inlinable_uniforms_valid_mask & BITFIELD64_BIT(pstage)) { 99 if (prog->inlined_variant_count[pstage] < ZINK_MAX_INLINED_VARIANTS) 100 base_size = zs->nir->info.num_inlinable_uniforms; 101 else 102 key->inline_uniforms = false; 103 } 104 105 struct zink_shader_module *iter, *next; 106 LIST_FOR_EACH_ENTRY_SAFE(iter, next, &prog->shader_cache[pstage][!!base_size], list) { 107 if (!shader_key_matches(iter, key, base_size)) 108 continue; 109 list_delinit(&iter->list); 110 zm = iter; 111 break; 112 } 113 114 if (!zm) { 115 zm = malloc(sizeof(struct zink_shader_module) + key->size + base_size * sizeof(uint32_t)); 116 if (!zm) { 117 return NULL; 118 } 119 mod = zink_shader_compile(screen, zs, prog->nir[stage], key); 120 if (!mod) { 121 FREE(zm); 122 return NULL; 123 } 124 zm->shader = mod; 125 list_inithead(&zm->list); 126 zm->num_uniforms = base_size; 127 zm->key_size = key->size; 128 memcpy(zm->key, key, key->size); 129 if (base_size) 130 memcpy(zm->key + key->size, &key->base, base_size * sizeof(uint32_t)); 131 zm->hash = shader_module_hash(zm); 132 zm->default_variant = !base_size && list_is_empty(&prog->shader_cache[pstage][0]); 133 if (base_size) 134 prog->inlined_variant_count[pstage]++; 135 } 136 list_add(&zm->list, &prog->shader_cache[pstage][!!base_size]); 137 return zm; 138} 139 140static void 141zink_destroy_shader_module(struct zink_screen *screen, struct zink_shader_module *zm) 142{ 143 VKSCR(DestroyShaderModule)(screen->dev, zm->shader, NULL); 144 free(zm); 145} 146 147static void 148destroy_shader_cache(struct zink_screen *screen, struct list_head *sc) 149{ 150 struct zink_shader_module *zm, *next; 151 LIST_FOR_EACH_ENTRY_SAFE(zm, next, sc, list) { 152 list_delinit(&zm->list); 153 zink_destroy_shader_module(screen, zm); 154 } 155} 156 157static void 158update_shader_modules(struct zink_context *ctx, 159 struct zink_screen *screen, 160 struct zink_gfx_program *prog, uint32_t mask, 161 struct zink_gfx_pipeline_state *state) 162{ 163 bool hash_changed = false; 164 bool default_variants = true; 165 bool first = !prog->modules[PIPE_SHADER_VERTEX]; 166 uint32_t variant_hash = prog->last_variant_hash; 167 u_foreach_bit(pstage, mask) { 168 assert(prog->shaders[pstage]); 169 struct zink_shader_module *zm = get_shader_module_for_stage(ctx, screen, prog->shaders[pstage], prog, state); 170 state->modules[pstage] = zm->shader; 171 if (prog->modules[pstage] == zm) 172 continue; 173 if (prog->modules[pstage]) 174 variant_hash ^= prog->modules[pstage]->hash; 175 hash_changed = true; 176 default_variants &= zm->default_variant; 177 prog->modules[pstage] = zm; 178 variant_hash ^= prog->modules[pstage]->hash; 179 } 180 181 if (hash_changed && state) { 182 if (default_variants && !first) 183 prog->last_variant_hash = prog->default_variant_hash; 184 else { 185 prog->last_variant_hash = variant_hash; 186 if (first) { 187 p_atomic_dec(&prog->base.reference.count); 188 prog->default_variant_hash = prog->last_variant_hash; 189 } 190 } 191 192 state->modules_changed = true; 193 } 194} 195 196static uint32_t 197hash_gfx_pipeline_state(const void *key) 198{ 199 const struct zink_gfx_pipeline_state *state = key; 200 uint32_t hash = _mesa_hash_data(key, offsetof(struct zink_gfx_pipeline_state, hash)); 201 if (!state->have_EXT_extended_dynamic_state2) 202 hash = XXH32(&state->primitive_restart, 1, hash); 203 if (state->have_EXT_extended_dynamic_state) 204 return hash; 205 return XXH32(&state->dyn_state1, sizeof(state->dyn_state1), hash); 206} 207 208static bool 209equals_gfx_pipeline_state(const void *a, const void *b) 210{ 211 const struct zink_gfx_pipeline_state *sa = a; 212 const struct zink_gfx_pipeline_state *sb = b; 213 if (!sa->have_EXT_extended_dynamic_state) { 214 if (sa->vertex_buffers_enabled_mask != sb->vertex_buffers_enabled_mask) 215 return false; 216 /* if we don't have dynamic states, we have to hash the enabled vertex buffer bindings */ 217 uint32_t mask_a = sa->vertex_buffers_enabled_mask; 218 uint32_t mask_b = sb->vertex_buffers_enabled_mask; 219 while (mask_a || mask_b) { 220 unsigned idx_a = u_bit_scan(&mask_a); 221 unsigned idx_b = u_bit_scan(&mask_b); 222 if (sa->vertex_strides[idx_a] != sb->vertex_strides[idx_b]) 223 return false; 224 } 225 if (sa->dyn_state1.front_face != sb->dyn_state1.front_face) 226 return false; 227 if (!!sa->dyn_state1.depth_stencil_alpha_state != !!sb->dyn_state1.depth_stencil_alpha_state || 228 (sa->dyn_state1.depth_stencil_alpha_state && 229 memcmp(sa->dyn_state1.depth_stencil_alpha_state, sb->dyn_state1.depth_stencil_alpha_state, 230 sizeof(struct zink_depth_stencil_alpha_hw_state)))) 231 return false; 232 } 233 if (!sa->have_EXT_extended_dynamic_state2) { 234 if (sa->primitive_restart != sb->primitive_restart) 235 return false; 236 } 237 return !memcmp(sa->modules, sb->modules, sizeof(sa->modules)) && 238 !memcmp(a, b, offsetof(struct zink_gfx_pipeline_state, hash)); 239} 240 241void 242zink_update_gfx_program(struct zink_context *ctx, struct zink_gfx_program *prog) 243{ 244 update_shader_modules(ctx, zink_screen(ctx->base.screen), prog, ctx->dirty_shader_stages & prog->stages_present, &ctx->gfx_pipeline_state); 245} 246 247VkPipelineLayout 248zink_pipeline_layout_create(struct zink_screen *screen, struct zink_program *pg, uint32_t *compat) 249{ 250 VkPipelineLayoutCreateInfo plci = {0}; 251 plci.sType = VK_STRUCTURE_TYPE_PIPELINE_LAYOUT_CREATE_INFO; 252 253 plci.pSetLayouts = pg->dsl; 254 plci.setLayoutCount = pg->num_dsl; 255 256 VkPushConstantRange pcr[2] = {0}; 257 if (pg->is_compute) { 258 if (((struct zink_compute_program*)pg)->shader->nir->info.stage == MESA_SHADER_KERNEL) { 259 pcr[0].stageFlags = VK_SHADER_STAGE_COMPUTE_BIT; 260 pcr[0].offset = 0; 261 pcr[0].size = sizeof(struct zink_cs_push_constant); 262 plci.pushConstantRangeCount = 1; 263 } 264 } else { 265 pcr[0].stageFlags = VK_SHADER_STAGE_VERTEX_BIT; 266 pcr[0].offset = offsetof(struct zink_gfx_push_constant, draw_mode_is_indexed); 267 pcr[0].size = 2 * sizeof(unsigned); 268 pcr[1].stageFlags = VK_SHADER_STAGE_TESSELLATION_CONTROL_BIT; 269 pcr[1].offset = offsetof(struct zink_gfx_push_constant, default_inner_level); 270 pcr[1].size = sizeof(float) * 6; 271 plci.pushConstantRangeCount = 2; 272 } 273 plci.pPushConstantRanges = &pcr[0]; 274 275 VkPipelineLayout layout; 276 if (VKSCR(CreatePipelineLayout)(screen->dev, &plci, NULL, &layout) != VK_SUCCESS) { 277 debug_printf("vkCreatePipelineLayout failed!\n"); 278 return VK_NULL_HANDLE; 279 } 280 281 *compat = _mesa_hash_data(pg->dsl, pg->num_dsl * sizeof(pg->dsl[0])); 282 283 return layout; 284} 285 286static void 287assign_io(struct zink_gfx_program *prog, struct zink_shader *stages[ZINK_SHADER_COUNT]) 288{ 289 struct zink_shader *shaders[PIPE_SHADER_TYPES]; 290 291 /* build array in pipeline order */ 292 for (unsigned i = 0; i < ZINK_SHADER_COUNT; i++) 293 shaders[tgsi_processor_to_shader_stage(i)] = stages[i]; 294 295 for (unsigned i = 0; i < MESA_SHADER_FRAGMENT;) { 296 nir_shader *producer = shaders[i]->nir; 297 for (unsigned j = i + 1; j < ZINK_SHADER_COUNT; i++, j++) { 298 struct zink_shader *consumer = shaders[j]; 299 if (!consumer) 300 continue; 301 if (!prog->nir[producer->info.stage]) 302 prog->nir[producer->info.stage] = nir_shader_clone(prog, producer); 303 if (!prog->nir[j]) 304 prog->nir[j] = nir_shader_clone(prog, consumer->nir); 305 zink_compiler_assign_io(prog->nir[producer->info.stage], prog->nir[j]); 306 i = j; 307 break; 308 } 309 } 310} 311 312struct zink_gfx_program * 313zink_create_gfx_program(struct zink_context *ctx, 314 struct zink_shader *stages[ZINK_SHADER_COUNT], 315 unsigned vertices_per_patch) 316{ 317 struct zink_screen *screen = zink_screen(ctx->base.screen); 318 struct zink_gfx_program *prog = rzalloc(NULL, struct zink_gfx_program); 319 if (!prog) 320 goto fail; 321 322 pipe_reference_init(&prog->base.reference, 1); 323 324 for (int i = 0; i < ZINK_SHADER_COUNT; ++i) { 325 list_inithead(&prog->shader_cache[i][0]); 326 list_inithead(&prog->shader_cache[i][1]); 327 if (stages[i]) { 328 prog->shaders[i] = stages[i]; 329 prog->stages_present |= BITFIELD_BIT(i); 330 } 331 } 332 if (stages[PIPE_SHADER_TESS_EVAL] && !stages[PIPE_SHADER_TESS_CTRL]) { 333 prog->shaders[PIPE_SHADER_TESS_EVAL]->generated = 334 prog->shaders[PIPE_SHADER_TESS_CTRL] = 335 zink_shader_tcs_create(screen, stages[PIPE_SHADER_VERTEX], vertices_per_patch); 336 prog->stages_present |= BITFIELD_BIT(PIPE_SHADER_TESS_CTRL); 337 } 338 339 assign_io(prog, prog->shaders); 340 341 if (stages[PIPE_SHADER_GEOMETRY]) 342 prog->last_vertex_stage = stages[PIPE_SHADER_GEOMETRY]; 343 else if (stages[PIPE_SHADER_TESS_EVAL]) 344 prog->last_vertex_stage = stages[PIPE_SHADER_TESS_EVAL]; 345 else 346 prog->last_vertex_stage = stages[PIPE_SHADER_VERTEX]; 347 348 for (int i = 0; i < ARRAY_SIZE(prog->pipelines); ++i) { 349 _mesa_hash_table_init(&prog->pipelines[i], prog, NULL, equals_gfx_pipeline_state); 350 /* only need first 3/4 for point/line/tri/patch */ 351 if (screen->info.have_EXT_extended_dynamic_state && 352 i == (prog->last_vertex_stage->nir->info.stage == MESA_SHADER_TESS_EVAL ? 4 : 3)) 353 break; 354 } 355 356 struct mesa_sha1 sctx; 357 _mesa_sha1_init(&sctx); 358 for (int i = 0; i < ZINK_SHADER_COUNT; ++i) { 359 if (prog->shaders[i]) { 360 simple_mtx_lock(&prog->shaders[i]->lock); 361 _mesa_set_add(prog->shaders[i]->programs, prog); 362 simple_mtx_unlock(&prog->shaders[i]->lock); 363 zink_gfx_program_reference(screen, NULL, prog); 364 _mesa_sha1_update(&sctx, prog->shaders[i]->base.sha1, sizeof(prog->shaders[i]->base.sha1)); 365 } 366 } 367 _mesa_sha1_final(&sctx, prog->base.sha1); 368 369 if (!screen->descriptor_program_init(ctx, &prog->base)) 370 goto fail; 371 372 zink_screen_get_pipeline_cache(screen, &prog->base); 373 return prog; 374 375fail: 376 if (prog) 377 zink_destroy_gfx_program(screen, prog); 378 return NULL; 379} 380 381static uint32_t 382hash_compute_pipeline_state(const void *key) 383{ 384 const struct zink_compute_pipeline_state *state = key; 385 uint32_t hash = _mesa_hash_data(state, offsetof(struct zink_compute_pipeline_state, hash)); 386 if (state->use_local_size) 387 hash = XXH32(&state->local_size[0], sizeof(state->local_size), hash); 388 return hash; 389} 390 391void 392zink_program_update_compute_pipeline_state(struct zink_context *ctx, struct zink_compute_program *comp, const uint block[3]) 393{ 394 struct zink_shader *zs = comp->shader; 395 bool use_local_size = !(zs->nir->info.workgroup_size[0] || 396 zs->nir->info.workgroup_size[1] || 397 zs->nir->info.workgroup_size[2]); 398 if (ctx->compute_pipeline_state.use_local_size != use_local_size) 399 ctx->compute_pipeline_state.dirty = true; 400 ctx->compute_pipeline_state.use_local_size = use_local_size; 401 402 if (ctx->compute_pipeline_state.use_local_size) { 403 for (int i = 0; i < ARRAY_SIZE(ctx->compute_pipeline_state.local_size); i++) { 404 if (ctx->compute_pipeline_state.local_size[i] != block[i]) 405 ctx->compute_pipeline_state.dirty = true; 406 ctx->compute_pipeline_state.local_size[i] = block[i]; 407 } 408 } else 409 ctx->compute_pipeline_state.local_size[0] = 410 ctx->compute_pipeline_state.local_size[1] = 411 ctx->compute_pipeline_state.local_size[2] = 0; 412} 413 414static bool 415equals_compute_pipeline_state(const void *a, const void *b) 416{ 417 return memcmp(a, b, offsetof(struct zink_compute_pipeline_state, hash)) == 0; 418} 419 420struct zink_compute_program * 421zink_create_compute_program(struct zink_context *ctx, struct zink_shader *shader) 422{ 423 struct zink_screen *screen = zink_screen(ctx->base.screen); 424 struct zink_compute_program *comp = rzalloc(NULL, struct zink_compute_program); 425 if (!comp) 426 goto fail; 427 428 pipe_reference_init(&comp->base.reference, 1); 429 comp->base.is_compute = true; 430 431 comp->module = CALLOC_STRUCT(zink_shader_module); 432 assert(comp->module); 433 comp->module->shader = zink_shader_compile(screen, shader, shader->nir, NULL); 434 assert(comp->module->shader); 435 436 comp->pipelines = _mesa_hash_table_create(NULL, hash_compute_pipeline_state, 437 equals_compute_pipeline_state); 438 439 _mesa_set_add(shader->programs, comp); 440 comp->shader = shader; 441 memcpy(comp->base.sha1, shader->base.sha1, sizeof(shader->base.sha1)); 442 443 if (!screen->descriptor_program_init(ctx, &comp->base)) 444 goto fail; 445 446 zink_screen_get_pipeline_cache(screen, &comp->base); 447 return comp; 448 449fail: 450 if (comp) 451 zink_destroy_compute_program(screen, comp); 452 return NULL; 453} 454 455uint32_t 456zink_program_get_descriptor_usage(struct zink_context *ctx, enum pipe_shader_type stage, enum zink_descriptor_type type) 457{ 458 struct zink_shader *zs = NULL; 459 switch (stage) { 460 case PIPE_SHADER_VERTEX: 461 case PIPE_SHADER_TESS_CTRL: 462 case PIPE_SHADER_TESS_EVAL: 463 case PIPE_SHADER_GEOMETRY: 464 case PIPE_SHADER_FRAGMENT: 465 zs = ctx->gfx_stages[stage]; 466 break; 467 case PIPE_SHADER_COMPUTE: { 468 zs = ctx->compute_stage; 469 break; 470 } 471 default: 472 unreachable("unknown shader type"); 473 } 474 if (!zs) 475 return 0; 476 switch (type) { 477 case ZINK_DESCRIPTOR_TYPE_UBO: 478 return zs->ubos_used; 479 case ZINK_DESCRIPTOR_TYPE_SSBO: 480 return zs->ssbos_used; 481 case ZINK_DESCRIPTOR_TYPE_SAMPLER_VIEW: 482 return BITSET_TEST_RANGE(zs->nir->info.textures_used, 0, PIPE_MAX_SAMPLERS - 1); 483 case ZINK_DESCRIPTOR_TYPE_IMAGE: 484 return zs->nir->info.images_used; 485 default: 486 unreachable("unknown descriptor type!"); 487 } 488 return 0; 489} 490 491bool 492zink_program_descriptor_is_buffer(struct zink_context *ctx, enum pipe_shader_type stage, enum zink_descriptor_type type, unsigned i) 493{ 494 struct zink_shader *zs = NULL; 495 switch (stage) { 496 case PIPE_SHADER_VERTEX: 497 case PIPE_SHADER_TESS_CTRL: 498 case PIPE_SHADER_TESS_EVAL: 499 case PIPE_SHADER_GEOMETRY: 500 case PIPE_SHADER_FRAGMENT: 501 zs = ctx->gfx_stages[stage]; 502 break; 503 case PIPE_SHADER_COMPUTE: { 504 zs = ctx->compute_stage; 505 break; 506 } 507 default: 508 unreachable("unknown shader type"); 509 } 510 if (!zs) 511 return false; 512 return zink_shader_descriptor_is_buffer(zs, type, i); 513} 514 515static unsigned 516get_num_bindings(struct zink_shader *zs, enum zink_descriptor_type type) 517{ 518 switch (type) { 519 case ZINK_DESCRIPTOR_TYPE_UBO: 520 case ZINK_DESCRIPTOR_TYPE_SSBO: 521 return zs->num_bindings[type]; 522 default: 523 break; 524 } 525 unsigned num_bindings = 0; 526 for (int i = 0; i < zs->num_bindings[type]; i++) 527 num_bindings += zs->bindings[type][i].size; 528 return num_bindings; 529} 530 531unsigned 532zink_program_num_bindings_typed(const struct zink_program *pg, enum zink_descriptor_type type, bool is_compute) 533{ 534 unsigned num_bindings = 0; 535 if (is_compute) { 536 struct zink_compute_program *comp = (void*)pg; 537 return get_num_bindings(comp->shader, type); 538 } 539 struct zink_gfx_program *prog = (void*)pg; 540 for (unsigned i = 0; i < ZINK_SHADER_COUNT; i++) { 541 if (prog->shaders[i]) 542 num_bindings += get_num_bindings(prog->shaders[i], type); 543 } 544 return num_bindings; 545} 546 547unsigned 548zink_program_num_bindings(const struct zink_program *pg, bool is_compute) 549{ 550 unsigned num_bindings = 0; 551 for (unsigned i = 0; i < ZINK_DESCRIPTOR_TYPES; i++) 552 num_bindings += zink_program_num_bindings_typed(pg, i, is_compute); 553 return num_bindings; 554} 555 556void 557zink_destroy_gfx_program(struct zink_screen *screen, 558 struct zink_gfx_program *prog) 559{ 560 if (prog->base.layout) 561 VKSCR(DestroyPipelineLayout)(screen->dev, prog->base.layout, NULL); 562 563 for (int i = 0; i < ZINK_SHADER_COUNT; ++i) { 564 if (prog->shaders[i]) { 565 _mesa_set_remove_key(prog->shaders[i]->programs, prog); 566 prog->shaders[i] = NULL; 567 } 568 destroy_shader_cache(screen, &prog->shader_cache[i][0]); 569 destroy_shader_cache(screen, &prog->shader_cache[i][1]); 570 ralloc_free(prog->nir[i]); 571 } 572 573 unsigned max_idx = ARRAY_SIZE(prog->pipelines); 574 if (screen->info.have_EXT_extended_dynamic_state) { 575 /* only need first 3/4 for point/line/tri/patch */ 576 if ((prog->stages_present & 577 (BITFIELD_BIT(PIPE_SHADER_TESS_EVAL) | BITFIELD_BIT(PIPE_SHADER_GEOMETRY))) == 578 BITFIELD_BIT(PIPE_SHADER_TESS_EVAL)) 579 max_idx = 4; 580 else 581 max_idx = 3; 582 max_idx++; 583 } 584 585 for (int i = 0; i < max_idx; ++i) { 586 hash_table_foreach(&prog->pipelines[i], entry) { 587 struct gfx_pipeline_cache_entry *pc_entry = entry->data; 588 589 VKSCR(DestroyPipeline)(screen->dev, pc_entry->pipeline, NULL); 590 free(pc_entry); 591 } 592 } 593 if (prog->base.pipeline_cache) 594 VKSCR(DestroyPipelineCache)(screen->dev, prog->base.pipeline_cache, NULL); 595 screen->descriptor_program_deinit(screen, &prog->base); 596 597 ralloc_free(prog); 598} 599 600void 601zink_destroy_compute_program(struct zink_screen *screen, 602 struct zink_compute_program *comp) 603{ 604 if (comp->base.layout) 605 VKSCR(DestroyPipelineLayout)(screen->dev, comp->base.layout, NULL); 606 607 if (comp->shader) 608 _mesa_set_remove_key(comp->shader->programs, comp); 609 610 hash_table_foreach(comp->pipelines, entry) { 611 struct compute_pipeline_cache_entry *pc_entry = entry->data; 612 613 VKSCR(DestroyPipeline)(screen->dev, pc_entry->pipeline, NULL); 614 free(pc_entry); 615 } 616 _mesa_hash_table_destroy(comp->pipelines, NULL); 617 VKSCR(DestroyShaderModule)(screen->dev, comp->module->shader, NULL); 618 free(comp->module); 619 if (comp->base.pipeline_cache) 620 VKSCR(DestroyPipelineCache)(screen->dev, comp->base.pipeline_cache, NULL); 621 screen->descriptor_program_deinit(screen, &comp->base); 622 623 ralloc_free(comp); 624} 625 626static unsigned 627get_pipeline_idx(bool have_EXT_extended_dynamic_state, enum pipe_prim_type mode, VkPrimitiveTopology vkmode) 628{ 629 /* VK_DYNAMIC_STATE_PRIMITIVE_TOPOLOGY_EXT specifies that the topology state in 630 * VkPipelineInputAssemblyStateCreateInfo only specifies the topology class, 631 * and the specific topology order and adjacency must be set dynamically 632 * with vkCmdSetPrimitiveTopologyEXT before any drawing commands. 633 */ 634 if (have_EXT_extended_dynamic_state) { 635 if (mode == PIPE_PRIM_PATCHES) 636 return 3; 637 switch (u_reduced_prim(mode)) { 638 case PIPE_PRIM_POINTS: 639 return 0; 640 case PIPE_PRIM_LINES: 641 return 1; 642 default: 643 return 2; 644 } 645 } 646 return vkmode; 647} 648 649 650VkPipeline 651zink_get_gfx_pipeline(struct zink_context *ctx, 652 struct zink_gfx_program *prog, 653 struct zink_gfx_pipeline_state *state, 654 enum pipe_prim_type mode) 655{ 656 struct zink_screen *screen = zink_screen(ctx->base.screen); 657 const bool have_EXT_vertex_input_dynamic_state = screen->info.have_EXT_vertex_input_dynamic_state; 658 const bool have_EXT_extended_dynamic_state = screen->info.have_EXT_extended_dynamic_state; 659 660 VkPrimitiveTopology vkmode = zink_primitive_topology(mode); 661 const unsigned idx = get_pipeline_idx(screen->info.have_EXT_extended_dynamic_state, mode, vkmode); 662 assert(idx <= ARRAY_SIZE(prog->pipelines)); 663 if (!state->dirty && !state->modules_changed && 664 (have_EXT_vertex_input_dynamic_state || !ctx->vertex_state_changed) && 665 idx == state->idx) 666 return state->pipeline; 667 668 struct hash_entry *entry = NULL; 669 670 if (state->dirty) { 671 if (state->pipeline) //avoid on first hash 672 state->final_hash ^= state->hash; 673 state->hash = hash_gfx_pipeline_state(state); 674 state->final_hash ^= state->hash; 675 state->dirty = false; 676 } 677 if (!have_EXT_vertex_input_dynamic_state && ctx->vertex_state_changed) { 678 if (state->pipeline) 679 state->final_hash ^= state->vertex_hash; 680 if (!have_EXT_extended_dynamic_state) { 681 uint32_t hash = 0; 682 /* if we don't have dynamic states, we have to hash the enabled vertex buffer bindings */ 683 uint32_t vertex_buffers_enabled_mask = state->vertex_buffers_enabled_mask; 684 hash = XXH32(&vertex_buffers_enabled_mask, sizeof(uint32_t), hash); 685 686 for (unsigned i = 0; i < state->element_state->num_bindings; i++) { 687 struct pipe_vertex_buffer *vb = ctx->vertex_buffers + ctx->element_state->binding_map[i]; 688 state->vertex_strides[i] = vb->buffer.resource ? vb->stride : 0; 689 hash = XXH32(&state->vertex_strides[i], sizeof(uint32_t), hash); 690 } 691 state->vertex_hash = hash ^ state->element_state->hash; 692 } else 693 state->vertex_hash = state->element_state->hash; 694 state->final_hash ^= state->vertex_hash; 695 } 696 state->modules_changed = false; 697 ctx->vertex_state_changed = false; 698 699 entry = _mesa_hash_table_search_pre_hashed(&prog->pipelines[idx], state->final_hash, state); 700 701 if (!entry) { 702 util_queue_fence_wait(&prog->base.cache_fence); 703 VkPipeline pipeline = zink_create_gfx_pipeline(screen, prog, 704 state, vkmode); 705 if (pipeline == VK_NULL_HANDLE) 706 return VK_NULL_HANDLE; 707 708 zink_screen_update_pipeline_cache(screen, &prog->base); 709 struct gfx_pipeline_cache_entry *pc_entry = CALLOC_STRUCT(gfx_pipeline_cache_entry); 710 if (!pc_entry) 711 return VK_NULL_HANDLE; 712 713 memcpy(&pc_entry->state, state, sizeof(*state)); 714 pc_entry->pipeline = pipeline; 715 716 entry = _mesa_hash_table_insert_pre_hashed(&prog->pipelines[idx], state->final_hash, pc_entry, pc_entry); 717 assert(entry); 718 } 719 720 struct gfx_pipeline_cache_entry *cache_entry = entry->data; 721 state->pipeline = cache_entry->pipeline; 722 state->idx = idx; 723 return state->pipeline; 724} 725 726VkPipeline 727zink_get_compute_pipeline(struct zink_screen *screen, 728 struct zink_compute_program *comp, 729 struct zink_compute_pipeline_state *state) 730{ 731 struct hash_entry *entry = NULL; 732 733 if (!state->dirty) 734 return state->pipeline; 735 if (state->dirty) { 736 state->hash = hash_compute_pipeline_state(state); 737 state->dirty = false; 738 } 739 entry = _mesa_hash_table_search_pre_hashed(comp->pipelines, state->hash, state); 740 741 if (!entry) { 742 util_queue_fence_wait(&comp->base.cache_fence); 743 VkPipeline pipeline = zink_create_compute_pipeline(screen, comp, state); 744 745 if (pipeline == VK_NULL_HANDLE) 746 return VK_NULL_HANDLE; 747 748 struct compute_pipeline_cache_entry *pc_entry = CALLOC_STRUCT(compute_pipeline_cache_entry); 749 if (!pc_entry) 750 return VK_NULL_HANDLE; 751 752 memcpy(&pc_entry->state, state, sizeof(*state)); 753 pc_entry->pipeline = pipeline; 754 755 entry = _mesa_hash_table_insert_pre_hashed(comp->pipelines, state->hash, pc_entry, pc_entry); 756 assert(entry); 757 } 758 759 struct compute_pipeline_cache_entry *cache_entry = entry->data; 760 state->pipeline = cache_entry->pipeline; 761 return state->pipeline; 762} 763 764static inline void 765bind_stage(struct zink_context *ctx, enum pipe_shader_type stage, 766 struct zink_shader *shader) 767{ 768 if (shader && shader->nir->info.num_inlinable_uniforms) 769 ctx->shader_has_inlinable_uniforms_mask |= 1 << stage; 770 else 771 ctx->shader_has_inlinable_uniforms_mask &= ~(1 << stage); 772 773 if (stage == PIPE_SHADER_COMPUTE) { 774 if (shader && shader != ctx->compute_stage) { 775 struct hash_entry *entry = _mesa_hash_table_search(&ctx->compute_program_cache, shader); 776 if (entry) { 777 ctx->compute_pipeline_state.dirty = true; 778 ctx->curr_compute = entry->data; 779 } else { 780 struct zink_compute_program *comp = zink_create_compute_program(ctx, shader); 781 _mesa_hash_table_insert(&ctx->compute_program_cache, comp->shader, comp); 782 ctx->compute_pipeline_state.dirty = true; 783 ctx->curr_compute = comp; 784 zink_batch_reference_program(&ctx->batch, &ctx->curr_compute->base); 785 } 786 } else if (!shader) 787 ctx->curr_compute = NULL; 788 ctx->compute_stage = shader; 789 zink_select_launch_grid(ctx); 790 } else { 791 if (ctx->gfx_stages[stage]) 792 ctx->gfx_hash ^= ctx->gfx_stages[stage]->hash; 793 ctx->gfx_stages[stage] = shader; 794 ctx->gfx_dirty = ctx->gfx_stages[PIPE_SHADER_FRAGMENT] && ctx->gfx_stages[PIPE_SHADER_VERTEX]; 795 ctx->gfx_pipeline_state.modules_changed = true; 796 if (shader) { 797 ctx->shader_stages |= BITFIELD_BIT(stage); 798 ctx->gfx_hash ^= ctx->gfx_stages[stage]->hash; 799 } else { 800 ctx->gfx_pipeline_state.modules[stage] = VK_NULL_HANDLE; 801 if (ctx->curr_program) 802 ctx->gfx_pipeline_state.final_hash ^= ctx->curr_program->last_variant_hash; 803 ctx->curr_program = NULL; 804 ctx->shader_stages &= ~BITFIELD_BIT(stage); 805 } 806 } 807} 808 809static void 810bind_last_vertex_stage(struct zink_context *ctx) 811{ 812 enum pipe_shader_type old = ctx->last_vertex_stage ? pipe_shader_type_from_mesa(ctx->last_vertex_stage->nir->info.stage) : PIPE_SHADER_TYPES; 813 if (ctx->gfx_stages[PIPE_SHADER_GEOMETRY]) 814 ctx->last_vertex_stage = ctx->gfx_stages[PIPE_SHADER_GEOMETRY]; 815 else if (ctx->gfx_stages[PIPE_SHADER_TESS_EVAL]) 816 ctx->last_vertex_stage = ctx->gfx_stages[PIPE_SHADER_TESS_EVAL]; 817 else 818 ctx->last_vertex_stage = ctx->gfx_stages[PIPE_SHADER_VERTEX]; 819 enum pipe_shader_type current = ctx->last_vertex_stage ? pipe_shader_type_from_mesa(ctx->last_vertex_stage->nir->info.stage) : PIPE_SHADER_VERTEX; 820 if (old != current) { 821 if (old != PIPE_SHADER_TYPES) { 822 memset(&ctx->gfx_pipeline_state.shader_keys.key[old].key.vs_base, 0, sizeof(struct zink_vs_key_base)); 823 ctx->dirty_shader_stages |= BITFIELD_BIT(old); 824 } else { 825 /* always unset vertex shader values when changing to a non-vs last stage */ 826 memset(&ctx->gfx_pipeline_state.shader_keys.key[PIPE_SHADER_VERTEX].key.vs_base, 0, sizeof(struct zink_vs_key_base)); 827 } 828 ctx->last_vertex_stage_dirty = true; 829 } 830} 831 832static void 833zink_bind_vs_state(struct pipe_context *pctx, 834 void *cso) 835{ 836 struct zink_context *ctx = zink_context(pctx); 837 if (!cso && !ctx->gfx_stages[PIPE_SHADER_VERTEX]) 838 return; 839 void *prev = ctx->gfx_stages[PIPE_SHADER_VERTEX]; 840 bind_stage(ctx, PIPE_SHADER_VERTEX, cso); 841 if (cso) { 842 struct zink_shader *zs = cso; 843 ctx->shader_reads_drawid = BITSET_TEST(zs->nir->info.system_values_read, SYSTEM_VALUE_DRAW_ID); 844 ctx->shader_reads_basevertex = BITSET_TEST(zs->nir->info.system_values_read, SYSTEM_VALUE_BASE_VERTEX); 845 } else { 846 ctx->shader_reads_drawid = false; 847 ctx->shader_reads_basevertex = false; 848 } 849 if (ctx->last_vertex_stage == prev) 850 ctx->last_vertex_stage = cso; 851 852} 853 854/* if gl_SampleMask[] is written to, we have to ensure that we get a shader with the same sample count: 855 * in GL, samples==1 means ignore gl_SampleMask[] 856 * in VK, gl_SampleMask[] is never ignored 857 */ 858void 859zink_update_fs_key_samples(struct zink_context *ctx) 860{ 861 if (!ctx->gfx_stages[PIPE_SHADER_FRAGMENT]) 862 return; 863 nir_shader *nir = ctx->gfx_stages[PIPE_SHADER_FRAGMENT]->nir; 864 if (nir->info.outputs_written & (1 << FRAG_RESULT_SAMPLE_MASK)) { 865 bool samples = zink_get_fs_key(ctx)->samples; 866 if (samples != (ctx->fb_state.samples > 1)) 867 zink_set_fs_key(ctx)->samples = ctx->fb_state.samples > 1; 868 } 869} 870 871static void 872zink_bind_fs_state(struct pipe_context *pctx, 873 void *cso) 874{ 875 struct zink_context *ctx = zink_context(pctx); 876 if (!cso && !ctx->gfx_stages[PIPE_SHADER_FRAGMENT]) 877 return; 878 bind_stage(ctx, PIPE_SHADER_FRAGMENT, cso); 879 ctx->fbfetch_outputs = 0; 880 if (cso) { 881 nir_shader *nir = ctx->gfx_stages[PIPE_SHADER_FRAGMENT]->nir; 882 if (nir->info.fs.uses_fbfetch_output) { 883 nir_foreach_shader_out_variable(var, ctx->gfx_stages[PIPE_SHADER_FRAGMENT]->nir) { 884 if (var->data.fb_fetch_output) 885 ctx->fbfetch_outputs |= BITFIELD_BIT(var->data.location - FRAG_RESULT_DATA0); 886 } 887 } 888 zink_update_fs_key_samples(ctx); 889 } 890 zink_update_fbfetch(ctx); 891} 892 893static void 894zink_bind_gs_state(struct pipe_context *pctx, 895 void *cso) 896{ 897 struct zink_context *ctx = zink_context(pctx); 898 if (!cso && !ctx->gfx_stages[PIPE_SHADER_GEOMETRY]) 899 return; 900 bool had_points = ctx->gfx_stages[PIPE_SHADER_GEOMETRY] ? ctx->gfx_stages[PIPE_SHADER_GEOMETRY]->nir->info.gs.output_primitive == GL_POINTS : false; 901 bind_stage(ctx, PIPE_SHADER_GEOMETRY, cso); 902 bind_last_vertex_stage(ctx); 903 if (cso) { 904 if (!had_points && ctx->last_vertex_stage->nir->info.gs.output_primitive == GL_POINTS) 905 ctx->gfx_pipeline_state.has_points++; 906 } else { 907 if (had_points) 908 ctx->gfx_pipeline_state.has_points--; 909 } 910} 911 912static void 913zink_bind_tcs_state(struct pipe_context *pctx, 914 void *cso) 915{ 916 bind_stage(zink_context(pctx), PIPE_SHADER_TESS_CTRL, cso); 917} 918 919static void 920zink_bind_tes_state(struct pipe_context *pctx, 921 void *cso) 922{ 923 struct zink_context *ctx = zink_context(pctx); 924 if (!cso && !ctx->gfx_stages[PIPE_SHADER_TESS_EVAL]) 925 return; 926 if (!!ctx->gfx_stages[PIPE_SHADER_TESS_EVAL] != !!cso) { 927 if (!cso) { 928 /* if unsetting a TESS that uses a generated TCS, ensure the TCS is unset */ 929 if (ctx->gfx_stages[PIPE_SHADER_TESS_EVAL]->generated) 930 ctx->gfx_stages[PIPE_SHADER_TESS_CTRL] = NULL; 931 } 932 } 933 bind_stage(ctx, PIPE_SHADER_TESS_EVAL, cso); 934 bind_last_vertex_stage(ctx); 935} 936 937static void * 938zink_create_cs_state(struct pipe_context *pctx, 939 const struct pipe_compute_state *shader) 940{ 941 struct nir_shader *nir; 942 if (shader->ir_type != PIPE_SHADER_IR_NIR) 943 nir = zink_tgsi_to_nir(pctx->screen, shader->prog); 944 else 945 nir = (struct nir_shader *)shader->prog; 946 947 return zink_shader_create(zink_screen(pctx->screen), nir, NULL); 948} 949 950static void 951zink_bind_cs_state(struct pipe_context *pctx, 952 void *cso) 953{ 954 bind_stage(zink_context(pctx), PIPE_SHADER_COMPUTE, cso); 955} 956 957void 958zink_delete_shader_state(struct pipe_context *pctx, void *cso) 959{ 960 zink_shader_free(zink_context(pctx), cso); 961} 962 963void * 964zink_create_gfx_shader_state(struct pipe_context *pctx, const struct pipe_shader_state *shader) 965{ 966 nir_shader *nir; 967 if (shader->type != PIPE_SHADER_IR_NIR) 968 nir = zink_tgsi_to_nir(pctx->screen, shader->tokens); 969 else 970 nir = (struct nir_shader *)shader->ir.nir; 971 972 return zink_shader_create(zink_screen(pctx->screen), nir, &shader->stream_output); 973} 974 975static void 976zink_delete_cached_shader_state(struct pipe_context *pctx, void *cso) 977{ 978 struct zink_screen *screen = zink_screen(pctx->screen); 979 util_shader_reference(pctx, &screen->shaders, &cso, NULL); 980} 981 982static void * 983zink_create_cached_shader_state(struct pipe_context *pctx, const struct pipe_shader_state *shader) 984{ 985 bool cache_hit; 986 struct zink_screen *screen = zink_screen(pctx->screen); 987 return util_live_shader_cache_get(pctx, &screen->shaders, shader, &cache_hit); 988} 989 990void 991zink_program_init(struct zink_context *ctx) 992{ 993 ctx->base.create_vs_state = zink_create_cached_shader_state; 994 ctx->base.bind_vs_state = zink_bind_vs_state; 995 ctx->base.delete_vs_state = zink_delete_cached_shader_state; 996 997 ctx->base.create_fs_state = zink_create_cached_shader_state; 998 ctx->base.bind_fs_state = zink_bind_fs_state; 999 ctx->base.delete_fs_state = zink_delete_cached_shader_state; 1000 1001 ctx->base.create_gs_state = zink_create_cached_shader_state; 1002 ctx->base.bind_gs_state = zink_bind_gs_state; 1003 ctx->base.delete_gs_state = zink_delete_cached_shader_state; 1004 1005 ctx->base.create_tcs_state = zink_create_cached_shader_state; 1006 ctx->base.bind_tcs_state = zink_bind_tcs_state; 1007 ctx->base.delete_tcs_state = zink_delete_cached_shader_state; 1008 1009 ctx->base.create_tes_state = zink_create_cached_shader_state; 1010 ctx->base.bind_tes_state = zink_bind_tes_state; 1011 ctx->base.delete_tes_state = zink_delete_cached_shader_state; 1012 1013 ctx->base.create_compute_state = zink_create_cs_state; 1014 ctx->base.bind_compute_state = zink_bind_cs_state; 1015 ctx->base.delete_compute_state = zink_delete_shader_state; 1016} 1017