/aosp_15_r20/external/mesa3d/src/intel/compiler/ |
H A D | brw_simd_selection.cpp | 87 const auto cs_prog_data = get_cs_prog_data(state); in brw_simd_should_compile() local 95 const bool workgroup_size_variable = cs_prog_data && cs_prog_data->local_size[0] == 0; in brw_simd_should_compile() 108 if (cs_prog_data) { in brw_simd_should_compile() 109 const unsigned workgroup_size = cs_prog_data->local_size[0] * in brw_simd_should_compile() 110 cs_prog_data->local_size[1] * in brw_simd_should_compile() 111 cs_prog_data->local_size[2]; in brw_simd_should_compile() 145 if (width == 32 && cs_prog_data && cs_prog_data->base.ray_queries > 0) { in brw_simd_should_compile() 150 if (width == 32 && cs_prog_data && cs_prog_data->uses_btd_stack_ids) { in brw_simd_should_compile() 200 auto cs_prog_data = get_cs_prog_data(state); in brw_simd_mark_compiled() local 203 if (cs_prog_data) in brw_simd_mark_compiled() [all …]
|
H A D | intel_clc.c | 148 const struct brw_cs_prog_data *cs_prog_data) in print_cs_prog_data_fields() argument 151 fprintf(fp, "%s." #field " = " fmt ",\n", pad, cs_prog_data->field) in print_cs_prog_data_fields() 155 cs_prog_data->field ? "true" : "false") in print_cs_prog_data_fields() 158 assert(cs_prog_data->base.stage == MESA_SHADER_COMPUTE); in print_cs_prog_data_fields() 160 assert(cs_prog_data->base.zero_push_reg == 0); in print_cs_prog_data_fields() 161 assert(cs_prog_data->base.push_reg_mask_param == 0); in print_cs_prog_data_fields() 172 assert(!cs_prog_data->base.has_ubo_pull); in print_cs_prog_data_fields() 173 assert(cs_prog_data->base.dispatch_grf_start_reg == 0); in print_cs_prog_data_fields() 174 assert(!cs_prog_data->base.use_alt_mode); in print_cs_prog_data_fields() 175 assert(cs_prog_data->base.param == 0); in print_cs_prog_data_fields() [all …]
|
H A D | brw_compile_cs.cpp | 31 struct brw_cs_prog_data *cs_prog_data) in cs_fill_push_const_info() argument 33 const struct brw_stage_prog_data *prog_data = &cs_prog_data->base; in cs_fill_push_const_info() 52 fill_push_const_block_info(&cs_prog_data->push.cross_thread, cross_thread_dwords); in cs_fill_push_const_info() 53 fill_push_const_block_info(&cs_prog_data->push.per_thread, per_thread_dwords); in cs_fill_push_const_info() 55 assert(cs_prog_data->push.cross_thread.dwords % 8 == 0 || in cs_fill_push_const_info() 56 cs_prog_data->push.per_thread.size == 0); in cs_fill_push_const_info() 57 assert(cs_prog_data->push.cross_thread.dwords + in cs_fill_push_const_info() 58 cs_prog_data->push.per_thread.dwords == in cs_fill_push_const_info()
|
H A D | brw_fs.cpp | 1793 brw_cs_push_const_total_size(const struct brw_cs_prog_data *cs_prog_data, in brw_cs_push_const_total_size() argument 1796 assert(cs_prog_data->push.per_thread.size % REG_SIZE == 0); in brw_cs_push_const_total_size() 1797 assert(cs_prog_data->push.cross_thread.size % REG_SIZE == 0); in brw_cs_push_const_total_size() 1798 return cs_prog_data->push.per_thread.size * threads + in brw_cs_push_const_total_size() 1799 cs_prog_data->push.cross_thread.size; in brw_cs_push_const_total_size()
|
H A D | brw_compiler.h | 1556 brw_cs_push_const_total_size(const struct brw_cs_prog_data *cs_prog_data,
|
H A D | brw_fs_nir.cpp | 4491 struct brw_cs_prog_data *cs_prog_data = brw_cs_prog_data(s.prog_data); in fs_nir_emit_cs_intrinsic() local 4513 cs_prog_data->uses_barrier = true; in fs_nir_emit_cs_intrinsic() 4523 assert(cs_prog_data->generate_local_id); in fs_nir_emit_cs_intrinsic() 4543 cs_prog_data->uses_num_work_groups = true; in fs_nir_emit_cs_intrinsic() 4595 cs_prog_data->uses_systolic = true; in fs_nir_emit_cs_intrinsic()
|
/aosp_15_r20/external/mesa3d/src/intel/compiler/elk/ |
H A D | elk_simd_selection.cpp | 77 const auto cs_prog_data = get_cs_prog_data(state); in elk_simd_should_compile() local 85 const bool workgroup_size_variable = cs_prog_data && cs_prog_data->local_size[0] == 0; in elk_simd_should_compile() 98 if (cs_prog_data) { in elk_simd_should_compile() 99 const unsigned workgroup_size = cs_prog_data->local_size[0] * in elk_simd_should_compile() 100 cs_prog_data->local_size[1] * in elk_simd_should_compile() 101 cs_prog_data->local_size[2]; in elk_simd_should_compile() 160 auto cs_prog_data = get_cs_prog_data(state); in elk_simd_mark_compiled() local 163 if (cs_prog_data) in elk_simd_mark_compiled() 164 cs_prog_data->prog_mask |= 1u << simd; in elk_simd_mark_compiled() 170 if (cs_prog_data) in elk_simd_mark_compiled() [all …]
|
H A D | elk_fs.cpp | 6811 elk_cs_push_const_total_size(const struct elk_cs_prog_data *cs_prog_data, in elk_cs_push_const_total_size() argument 6814 assert(cs_prog_data->push.per_thread.size % REG_SIZE == 0); in elk_cs_push_const_total_size() 6815 assert(cs_prog_data->push.cross_thread.size % REG_SIZE == 0); in elk_cs_push_const_total_size() 6816 return cs_prog_data->push.per_thread.size * threads + in elk_cs_push_const_total_size() 6817 cs_prog_data->push.cross_thread.size; in elk_cs_push_const_total_size() 6830 struct elk_cs_prog_data *cs_prog_data) in cs_fill_push_const_info() argument 6832 const struct elk_stage_prog_data *prog_data = &cs_prog_data->base; in cs_fill_push_const_info() 6855 fill_push_const_block_info(&cs_prog_data->push.cross_thread, cross_thread_dwords); in cs_fill_push_const_info() 6856 fill_push_const_block_info(&cs_prog_data->push.per_thread, per_thread_dwords); in cs_fill_push_const_info() 6858 assert(cs_prog_data->push.cross_thread.dwords % 8 == 0 || in cs_fill_push_const_info() [all …]
|
H A D | elk_compiler.h | 1658 elk_cs_push_const_total_size(const struct elk_cs_prog_data *cs_prog_data,
|
H A D | elk_fs_nir.cpp | 4004 struct elk_cs_prog_data *cs_prog_data = elk_cs_prog_data(s.prog_data); in fs_nir_emit_cs_intrinsic() local 4026 cs_prog_data->uses_barrier = true; in fs_nir_emit_cs_intrinsic() 4046 cs_prog_data->uses_num_work_groups = true; in fs_nir_emit_cs_intrinsic()
|
/aosp_15_r20/external/mesa3d/src/intel/blorp/ |
H A D | blorp_genX_exec_brw.h | 1579 const struct brw_cs_prog_data *cs_prog_data = params->cs_prog_data; in blorp_get_compute_push_const() local 1581 ALIGN(brw_cs_push_const_total_size(cs_prog_data, threads), 64); in blorp_get_compute_push_const() 1582 assert(cs_prog_data->push.cross_thread.size + in blorp_get_compute_push_const() 1583 cs_prog_data->push.per_thread.size == sizeof(params->wm_inputs)); in blorp_get_compute_push_const() 1608 if (cs_prog_data->push.cross_thread.size > 0) { in blorp_get_compute_push_const() 1609 memcpy(dst, src, cs_prog_data->push.cross_thread.size); in blorp_get_compute_push_const() 1610 dst += cs_prog_data->push.cross_thread.size; in blorp_get_compute_push_const() 1611 src += cs_prog_data->push.cross_thread.size; in blorp_get_compute_push_const() 1614 assert(GFX_VERx10 < 125 || cs_prog_data->push.per_thread.size == 0); in blorp_get_compute_push_const() 1616 if (cs_prog_data->push.per_thread.size > 0) { in blorp_get_compute_push_const() [all …]
|
H A D | blorp_brw.c | 136 struct brw_cs_prog_data *cs_prog_data = rzalloc(mem_ctx, struct brw_cs_prog_data); in blorp_compile_cs_brw() local 137 cs_prog_data->base.nr_params = nr_params; in blorp_compile_cs_brw() 138 cs_prog_data->base.param = rzalloc_array(NULL, uint32_t, nr_params); in blorp_compile_cs_brw() 141 cs_prog_data); in blorp_compile_cs_brw() 156 .prog_data = cs_prog_data, in blorp_compile_cs_brw() 161 ralloc_free(cs_prog_data->base.param); in blorp_compile_cs_brw() 162 cs_prog_data->base.param = NULL; in blorp_compile_cs_brw() 166 .kernel_size = cs_prog_data->base.program_size, in blorp_compile_cs_brw() 167 .prog_data = cs_prog_data, in blorp_compile_cs_brw() 168 .prog_data_size = sizeof(*cs_prog_data), in blorp_compile_cs_brw()
|
H A D | blorp_elk.c | 144 struct elk_cs_prog_data *cs_prog_data = rzalloc(mem_ctx, struct elk_cs_prog_data); in blorp_compile_cs_elk() local 145 cs_prog_data->base.nr_params = nr_params; in blorp_compile_cs_elk() 146 cs_prog_data->base.param = rzalloc_array(NULL, uint32_t, nr_params); in blorp_compile_cs_elk() 149 cs_prog_data); in blorp_compile_cs_elk() 164 .prog_data = cs_prog_data, in blorp_compile_cs_elk() 169 ralloc_free(cs_prog_data->base.param); in blorp_compile_cs_elk() 170 cs_prog_data->base.param = NULL; in blorp_compile_cs_elk() 174 .kernel_size = cs_prog_data->base.program_size, in blorp_compile_cs_elk() 175 .prog_data = cs_prog_data, in blorp_compile_cs_elk() 176 .prog_data_size = sizeof(*cs_prog_data), in blorp_compile_cs_elk()
|
H A D | blorp_genX_exec_elk.h | 1887 const struct elk_cs_prog_data *cs_prog_data = params->cs_prog_data; in blorp_get_compute_push_const() local 1889 ALIGN(elk_cs_push_const_total_size(cs_prog_data, threads), 64); in blorp_get_compute_push_const() 1890 assert(cs_prog_data->push.cross_thread.size + in blorp_get_compute_push_const() 1891 cs_prog_data->push.per_thread.size == sizeof(params->wm_inputs)); in blorp_get_compute_push_const() 1908 if (cs_prog_data->push.cross_thread.size > 0) { in blorp_get_compute_push_const() 1909 memcpy(dst, src, cs_prog_data->push.cross_thread.size); in blorp_get_compute_push_const() 1910 dst += cs_prog_data->push.cross_thread.size; in blorp_get_compute_push_const() 1911 src += cs_prog_data->push.cross_thread.size; in blorp_get_compute_push_const() 1914 if (cs_prog_data->push.per_thread.size > 0) { in blorp_get_compute_push_const() 1916 memcpy(dst, src, (cs_prog_data->push.per_thread.dwords - 1) * 4); in blorp_get_compute_push_const() [all …]
|
H A D | blorp_priv.h | 266 void *cs_prog_data; member
|
H A D | blorp_clear.c | 130 ¶ms->cs_prog_kernel, ¶ms->cs_prog_data)) in blorp_params_get_clear_kernel_cs() 177 ¶ms->cs_prog_kernel, ¶ms->cs_prog_data); in blorp_params_get_clear_kernel_cs()
|
H A D | blorp_blit.c | 1538 ¶ms->cs_prog_kernel, ¶ms->cs_prog_data)) in blorp_get_blit_kernel_cs() 1558 ¶ms->cs_prog_kernel, ¶ms->cs_prog_data); in blorp_get_blit_kernel_cs()
|
/aosp_15_r20/external/mesa3d/src/intel/vulkan_hasvk/ |
H A D | anv_cmd_buffer.c | 579 const struct elk_cs_prog_data *cs_prog_data = get_cs_prog_data(pipeline); in anv_cmd_buffer_cs_push_constants() local 583 elk_cs_get_dispatch_info(devinfo, cs_prog_data, NULL); in anv_cmd_buffer_cs_push_constants() 585 elk_cs_push_const_total_size(cs_prog_data, dispatch.threads); in anv_cmd_buffer_cs_push_constants() 601 if (cs_prog_data->push.cross_thread.size > 0) { in anv_cmd_buffer_cs_push_constants() 602 memcpy(dst, src, cs_prog_data->push.cross_thread.size); in anv_cmd_buffer_cs_push_constants() 603 dst += cs_prog_data->push.cross_thread.size; in anv_cmd_buffer_cs_push_constants() 604 src += cs_prog_data->push.cross_thread.size; in anv_cmd_buffer_cs_push_constants() 607 if (cs_prog_data->push.per_thread.size > 0) { in anv_cmd_buffer_cs_push_constants() 609 memcpy(dst, src, cs_prog_data->push.per_thread.size); in anv_cmd_buffer_cs_push_constants() 613 (range->start * 32 + cs_prog_data->push.cross_thread.size); in anv_cmd_buffer_cs_push_constants() [all …]
|
H A D | genX_pipeline.c | 1882 const struct elk_cs_prog_data *cs_prog_data = get_cs_prog_data(pipeline); in genX() local 1884 anv_pipeline_setup_l3_config(&pipeline->base, cs_prog_data->base.total_shared > 0); in genX() 1887 elk_cs_get_dispatch_info(devinfo, cs_prog_data, NULL); in genX() 1889 ALIGN(cs_prog_data->push.per_thread.regs * dispatch.threads + in genX() 1890 cs_prog_data->push.cross_thread.regs, 2); in genX() 1936 elk_cs_prog_data_prog_offset(cs_prog_data, dispatch.simd_size), in genX() 1942 .BarrierEnable = cs_prog_data->uses_barrier, in genX() 1943 … .SharedLocalMemorySize = intel_compute_slm_encode_size(GFX_VER, cs_prog_data->base.total_shared), in genX() 1948 .ConstantURBEntryReadLength = cs_prog_data->push.per_thread.regs, in genX() 1951 cs_prog_data->push.cross_thread.regs, in genX()
|
/aosp_15_r20/external/mesa3d/src/intel/vulkan/ |
H A D | anv_cmd_buffer.c | 1239 const struct brw_cs_prog_data *cs_prog_data = get_cs_prog_data(pipeline); in anv_cmd_buffer_cs_push_constants() local 1243 brw_cs_get_dispatch_info(devinfo, cs_prog_data, NULL); in anv_cmd_buffer_cs_push_constants() 1245 brw_cs_push_const_total_size(cs_prog_data, dispatch.threads); in anv_cmd_buffer_cs_push_constants() 1268 if (cs_prog_data->push.cross_thread.size > 0) { in anv_cmd_buffer_cs_push_constants() 1269 memcpy(dst, src, cs_prog_data->push.cross_thread.size); in anv_cmd_buffer_cs_push_constants() 1270 dst += cs_prog_data->push.cross_thread.size; in anv_cmd_buffer_cs_push_constants() 1271 src += cs_prog_data->push.cross_thread.size; in anv_cmd_buffer_cs_push_constants() 1274 if (cs_prog_data->push.per_thread.size > 0) { in anv_cmd_buffer_cs_push_constants() 1276 memcpy(dst, src, cs_prog_data->push.per_thread.size); in anv_cmd_buffer_cs_push_constants() 1280 (range->start * 32 + cs_prog_data->push.cross_thread.size); in anv_cmd_buffer_cs_push_constants() [all …]
|
H A D | genX_pipeline.c | 2148 const struct brw_cs_prog_data *cs_prog_data = get_cs_prog_data(pipeline); in genX() local 2149 anv_pipeline_setup_l3_config(&pipeline->base, cs_prog_data->base.total_shared > 0); in genX() 2159 const struct brw_cs_prog_data *cs_prog_data = get_cs_prog_data(pipeline); in genX() local 2161 anv_pipeline_setup_l3_config(&pipeline->base, cs_prog_data->base.total_shared > 0); in genX() 2164 brw_cs_get_dispatch_info(devinfo, cs_prog_data, NULL); in genX() 2166 ALIGN(cs_prog_data->push.per_thread.regs * dispatch.threads + in genX() 2167 cs_prog_data->push.cross_thread.regs, 2); in genX() 2182 if (cs_prog_data->base.total_scratch) { in genX() 2186 vfe.PerThreadScratchSpace = ffs(cs_prog_data->base.total_scratch) - 11; in genX() 2195 brw_cs_prog_data_prog_offset(cs_prog_data, dispatch.simd_size), in genX() [all …]
|
H A D | genX_cmd_compute.c | 601 const struct brw_cs_prog_data *cs_prog_data = in genX() local 653 brw_cs_get_dispatch_info(devinfo, cs_prog_data, NULL); in genX() 661 cw.LocalXMaximum = cs_prog_data->local_size[0] - 1; in genX() 662 cw.LocalYMaximum = cs_prog_data->local_size[1] - 1; in genX() 663 cw.LocalZMaximum = cs_prog_data->local_size[2] - 1; in genX() 678 cs_prog_data, in genX() 1058 const struct brw_cs_prog_data *cs_prog_data = in cmd_buffer_trace_rays() local 1061 brw_cs_get_dispatch_info(device->info, cs_prog_data, NULL); in cmd_buffer_trace_rays()
|
/aosp_15_r20/external/mesa3d/src/gallium/drivers/crocus/ |
H A D | crocus_program.c | 2505 struct elk_cs_prog_data *cs_prog_data = in crocus_compile_cs() local 2507 struct elk_stage_prog_data *prog_data = &cs_prog_data->base; in crocus_compile_cs() 2515 NIR_PASS_V(nir, elk_nir_lower_cs_intrinsics, devinfo, cs_prog_data); in crocus_compile_cs() 2531 .prog_data = cs_prog_data, in crocus_compile_cs() 2551 prog_data, sizeof(*cs_prog_data), NULL, in crocus_compile_cs() 2608 crocus_fill_cs_push_const_buffer(struct elk_cs_prog_data *cs_prog_data, in crocus_fill_cs_push_const_buffer() argument 2612 assert(elk_cs_push_const_total_size(cs_prog_data, threads) > 0); in crocus_fill_cs_push_const_buffer() 2613 assert(cs_prog_data->push.cross_thread.size == 0); in crocus_fill_cs_push_const_buffer() 2614 assert(cs_prog_data->push.per_thread.dwords == 1); in crocus_fill_cs_push_const_buffer() 2615 assert(cs_prog_data->base.param[0] == ELK_PARAM_BUILTIN_SUBGROUP_ID); in crocus_fill_cs_push_const_buffer()
|
H A D | crocus_state.c | 8054 struct elk_cs_prog_data *cs_prog_data = (void *) prog_data; local 8056 elk_cs_get_dispatch_info(devinfo, cs_prog_data, grid->block); 8074 cs_prog_data->local_size[0] == 0 /* Variable local group size */) { 8126 ALIGN(cs_prog_data->push.per_thread.regs * dispatch.threads + 8127 cs_prog_data->push.cross_thread.regs, 2); 8133 cs_prog_data->local_size[0] == 0 /* Variable local group size */) { 8135 assert(cs_prog_data->push.cross_thread.dwords == 0 && 8136 cs_prog_data->push.per_thread.dwords == 1 && 8137 cs_prog_data->base.param[0] == ELK_PARAM_BUILTIN_SUBGROUP_ID); 8139 elk_cs_push_const_total_size(cs_prog_data, dispatch.threads); [all …]
|
H A D | crocus_context.h | 776 void crocus_fill_cs_push_const_buffer(struct elk_cs_prog_data *cs_prog_data,
|