| /external/mesa3d/src/intel/compiler/ |
| D | brw_fs_builder.h | 56 unsigned dispatch_width) : in fs_builder() argument
58 _dispatch_width(dispatch_width), in fs_builder()
119 if (n <= dispatch_width() && i < dispatch_width() / n) { in group()
178 dispatch_width() const in dispatch_width() function
201 assert(dispatch_width() <= 32);
205 DIV_ROUND_UP(n * type_sz(type) * dispatch_width(),
260 return emit(instruction(opcode, dispatch_width())); in emit()
269 return emit(instruction(opcode, dispatch_width(), dst)); in emit()
286 return emit(instruction(opcode, dispatch_width(), dst, in emit()
290 return emit(instruction(opcode, dispatch_width(), dst, src0)); in emit()
[all …]
|
| D | brw_vec4_builder.h | 54 vec4_builder(backend_shader *shader, unsigned dispatch_width = 8) :
56 _dispatch_width(dispatch_width), _group(0), in shader()
114 (n <= dispatch_width() && i < dispatch_width() / n)); in group()
151 dispatch_width() const in dispatch_width() function
174 assert(dispatch_width() <= 32);
190 return dst_reg(retype(brw_null_vec(dispatch_width()), in null_reg_f()
200 return dst_reg(retype(brw_null_vec(dispatch_width()), in null_reg_d()
210 return dst_reg(retype(brw_null_vec(dispatch_width()), in null_reg_ud()
313 inst->exec_size = dispatch_width(); in emit()
|
| D | brw_fs_reg_allocate.cpp | 50 int reg_width = dispatch_width / 8; in assign_regs_trivial()
80 aligned_bary_size(unsigned dispatch_width) in aligned_bary_size() argument
82 return (dispatch_width == 8 ? 2 : 4); in aligned_bary_size()
86 brw_alloc_reg_set(struct brw_compiler *compiler, int dispatch_width) in brw_alloc_reg_set() argument
90 const int index = util_logbase2(dispatch_width / 8); in brw_alloc_reg_set()
92 if (dispatch_width > 8 && devinfo->gen >= 7) { in brw_alloc_reg_set()
128 if (devinfo->gen <= 5 && dispatch_width >= 16) { in brw_alloc_reg_set()
176 if (devinfo->gen <= 5 && dispatch_width >= 16) { in brw_alloc_reg_set()
217 if (class_sizes[i] == int(aligned_bary_size(dispatch_width))) { in brw_alloc_reg_set()
222 if (devinfo->gen <= 5 && dispatch_width >= 16) { in brw_alloc_reg_set()
[all …]
|
| D | brw_fs_visitor.cpp | 96 int reg_width = dispatch_width / 8; in emit_dummy_fs()
181 for (unsigned i = 0; i < dispatch_width / 8; i++) { in emit_interpolation_setup_gen4()
277 for (unsigned i = 0; i < DIV_ROUND_UP(dispatch_width, 16); i++) { in emit_interpolation_setup_gen6()
278 const fs_builder hbld = abld.group(MIN2(16, dispatch_width), i); in emit_interpolation_setup_gen6()
281 if (devinfo->gen >= 8 || dispatch_width == 8) { in emit_interpolation_setup_gen6()
293 abld.exec_all().group(hbld.dispatch_width() * 2, 0); in emit_interpolation_setup_gen6()
351 for (unsigned i = 0; i < DIV_ROUND_UP(dispatch_width, 16); i++) { in emit_interpolation_setup_gen6()
367 for (unsigned q = 0; q < dispatch_width / 8; q++) { in emit_interpolation_setup_gen6()
905 unsigned dispatch_width, in fs_visitor() argument
913 dispatch_width(dispatch_width), in fs_visitor()
[all …]
|
| D | brw_fs.h | 78 return offset(reg, bld.dispatch_width(), delta); in offset()
101 unsigned dispatch_width,
419 const unsigned dispatch_width; /**< 8, 16 or 32 */ variable
472 int generate_code(const cfg_t *cfg, int dispatch_width,
572 unsigned dispatch_width; /**< 8, 16 or 32 */ variable
590 if (bld.dispatch_width() > 16) {
593 const unsigned m = bld.dispatch_width() / hbld.dispatch_width();
617 const unsigned m = bld.dispatch_width() / hbld.dispatch_width(); in fetch_barycentric_reg()
|
| D | brw_compiler.h | 1004 unsigned dispatch_width) in brw_cs_prog_data_prog_offset() argument
1006 assert(dispatch_width == 8 || in brw_cs_prog_data_prog_offset()
1007 dispatch_width == 16 || in brw_cs_prog_data_prog_offset()
1008 dispatch_width == 32); in brw_cs_prog_data_prog_offset()
1009 const unsigned index = dispatch_width / 16; in brw_cs_prog_data_prog_offset()
1367 uint32_t dispatch_width; /**< 0 for vec4 */ member
|
| D | brw_fs_nir.cpp | 45 last_scratch = ALIGN(nir->scratch_size, 4) * dispatch_width; in emit_nir_code()
218 for (unsigned i = 0; i < DIV_ROUND_UP(v->dispatch_width, 16); i++) { in emit_system_values_block()
219 const fs_builder hbld = abld.group(MIN2(16, v->dispatch_width), i); in emit_system_values_block()
279 if (dispatch_width > 8) in nir_emit_system_values()
281 if (dispatch_width > 16) { in nir_emit_system_values()
3749 workgroup_size() <= dispatch_width) { in nir_emit_cs_intrinsic()
3789 inst->size_written = 3 * dispatch_width * 4; in nir_emit_cs_intrinsic()
3835 inst->size_written = instr->num_components * dispatch_width * 4; in nir_emit_cs_intrinsic()
4046 const unsigned chan_index_bits = ffs(dispatch_width) - 1; in swizzle_nir_scratch_addr()
4174 inst->size_written = instr->num_components * dispatch_width * 4; in nir_emit_intrinsic()
[all …]
|
| D | brw_fs.cpp | 661 if (dispatch_width == 8) in SHADER_TIME_ADD()
681 dispatch_width, stage_abbrev, msg); in vfail()
714 if (dispatch_width > n) { in limit_dispatch_width()
1189 int reg_width = dispatch_width / 8; in vgrf()
1446 for (unsigned i = 0; i < DIV_ROUND_UP(dispatch_width, 16); i++) { in emit_sampleid_setup()
1447 const fs_builder hbld = abld.group(MIN2(16, dispatch_width), i); in emit_sampleid_setup()
3399 if (dispatch_width >= 16) in remove_duplicate_mrf_writes()
3868 fs_reg imm(VGRF, alloc.allocate(dispatch_width / 8), inst->dst.type); in lower_mul_dword_inst()
4297 assert(bld.dispatch_width() <= 16); in sample_mask_reg()
4300 assert(v->devinfo->gen >= 6 && bld.dispatch_width() <= 16); in sample_mask_reg()
[all …]
|
| D | brw_wm_iz.cpp | 125 assert(dispatch_width <= 16); in setup_fs_payload_gen4()
|
| D | brw_fs_generator.cpp | 194 prog_data(prog_data), dispatch_width(0), in fs_generator()
491 inst->exec_size == dispatch_width; in generate_mov_indirect()
671 lower_width == dispatch_width; in generate_shuffle()
1074 assert(!inst->eot || inst->exec_size == dispatch_width); in generate_tex()
1874 fs_generator::generate_code(const cfg_t *cfg, int dispatch_width, in generate_code() argument
1882 this->dispatch_width = dispatch_width; in generate_code()
2656 dispatch_width, before_size / 16, in generate_code()
2688 dispatch_width, before_size / 16 - nop_count, in generate_code()
2695 stats->dispatch_width = dispatch_width; in generate_code()
|
| D | brw_ir_performance.cpp | 1506 unsigned dispatch_width) in calculate_performance() argument
1533 const float discard_weight = (dispatch_width > 16 || s->devinfo->gen < 12 ? in calculate_performance()
1565 p.throughput = dispatch_width * calculate_thread_throughput(st, elapsed); in calculate_performance()
1572 calculate_performance(*this, v, issue_fs_inst, v->dispatch_width); in performance()
|
| D | brw_vec4_generator.cpp | 2217 stats->dispatch_width = 0; in generate_code()
|
| /external/igt-gpu-tools/assembler/ |
| D | brw_eu.h | 251 int dispatch_width,
|
| D | brw_eu_emit.c | 2132 int dispatch_width, in brw_fb_WRITE() argument
2147 if (dispatch_width == 16) in brw_fb_WRITE()
|
| /external/mesa3d/src/intel/vulkan/ |
| D | anv_pipeline.c | 2274 unsigned simd_width = exe->stats.dispatch_width; in anv_GetPipelineExecutablePropertiesKHR()
|