| /external/mesa3d/src/compiler/nir/ |
| D | nir_lower_system_values.c | 60 nir_const_value local_size; in convert_block() local
61 memset(&local_size, 0, sizeof(local_size)); in convert_block()
62 local_size.u32[0] = b->shader->info.cs.local_size[0]; in convert_block()
63 local_size.u32[1] = b->shader->info.cs.local_size[1]; in convert_block()
64 local_size.u32[2] = b->shader->info.cs.local_size[2]; in convert_block()
70 nir_build_imm(b, 3, 32, local_size)), in convert_block()
92 nir_imm_int(b, b->shader->info.cs.local_size[0]); in convert_block()
94 nir_imm_int(b, b->shader->info.cs.local_size[1]); in convert_block()
|
| /external/pcre/dist2/src/sljit/ |
| D | sljitNativeX86_32.c | 68 sljit_s32 fscratches, sljit_s32 fsaveds, sljit_s32 local_size) in sljit_emit_enter() argument
74 …ljit_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size)); in sljit_emit_enter()
75 set_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size); in sljit_emit_enter()
166 local_size = ((SLJIT_LOCALS_OFFSET + saveds + local_size + 15) & ~15) - saveds; in sljit_emit_enter()
169 …local_size = SLJIT_LOCALS_OFFSET + ((local_size + sizeof(sljit_f64) - 1) & ~(sizeof(sljit_f64) - 1… in sljit_emit_enter()
171 …local_size = SLJIT_LOCALS_OFFSET + ((local_size + sizeof(sljit_sw) - 1) & ~(sizeof(sljit_sw) - 1)); in sljit_emit_enter()
174 compiler->local_size = local_size; in sljit_emit_enter()
177 if (local_size > 0) { in sljit_emit_enter()
178 if (local_size <= 4 * 4096) { in sljit_emit_enter()
179 if (local_size > 4096) in sljit_emit_enter()
[all …]
|
| D | sljitNativeX86_64.c | 77 sljit_s32 fscratches, sljit_s32 fsaveds, sljit_s32 local_size) in sljit_emit_enter() argument
83 …ljit_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size)); in sljit_emit_enter()
84 set_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size); in sljit_emit_enter()
168 …local_size = ((local_size + SLJIT_LOCALS_OFFSET + saved_register_size + 15) & ~15) - saved_registe… in sljit_emit_enter()
169 compiler->local_size = local_size; in sljit_emit_enter()
172 if (local_size > 0) { in sljit_emit_enter()
173 if (local_size <= 4 * 4096) { in sljit_emit_enter()
174 if (local_size > 4096) in sljit_emit_enter()
176 if (local_size > 2 * 4096) in sljit_emit_enter()
178 if (local_size > 3 * 4096) in sljit_emit_enter()
[all …]
|
| D | sljitNativeARM_64.c | 870 sljit_s32 fscratches, sljit_s32 fsaveds, sljit_s32 local_size) in sljit_emit_enter() argument
875 …ljit_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size)); in sljit_emit_enter()
876 set_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size); in sljit_emit_enter()
882 local_size = (local_size + 15) & ~0xf; in sljit_emit_enter()
883 compiler->local_size = local_size + saved_regs_size; in sljit_emit_enter()
889 if (local_size >= 4096) in sljit_emit_enter()
891 else if (local_size > 256) in sljit_emit_enter()
892 FAIL_IF(push_inst(compiler, SUBI | RD(TMP_REG1) | RN(SLJIT_SP) | (local_size << 10))); in sljit_emit_enter()
934 if (local_size >= 4096) { in sljit_emit_enter()
935 if (local_size < 4 * 4096) { in sljit_emit_enter()
[all …]
|
| D | sljitNativeARM_T2_32.c | 1037 sljit_s32 fscratches, sljit_s32 fsaveds, sljit_s32 local_size) in sljit_emit_enter() argument
1046 …ljit_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size)); in sljit_emit_enter()
1047 set_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size); in sljit_emit_enter()
1062 local_size = ((size + local_size + 7) & ~7) - size; in sljit_emit_enter()
1063 compiler->local_size = local_size; in sljit_emit_enter()
1066 if (local_size >= 256) { in sljit_emit_enter()
1067 if (local_size > 4096) in sljit_emit_enter()
1070 imm = get_imm(local_size & ~0xff); in sljit_emit_enter()
1076 if (local_size > 0) { in sljit_emit_enter()
1077 if (local_size <= (127 << 2)) in sljit_emit_enter()
[all …]
|
| D | sljitNativeMIPS_common.c | 594 sljit_s32 fscratches, sljit_s32 fsaveds, sljit_s32 local_size) in sljit_emit_enter() argument
600 …ljit_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size)); in sljit_emit_enter()
601 set_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size); in sljit_emit_enter()
603 local_size += GET_SAVED_REGISTERS_SIZE(scratches, saveds, 1) + SLJIT_LOCALS_OFFSET; in sljit_emit_enter()
605 local_size = (local_size + 15) & ~0xf; in sljit_emit_enter()
607 local_size = (local_size + 31) & ~0x1f; in sljit_emit_enter()
609 compiler->local_size = local_size; in sljit_emit_enter()
611 if (local_size <= SIMM_MAX) { in sljit_emit_enter()
613 …FAIL_IF(push_inst(compiler, ADDIU_W | S(SLJIT_SP) | T(SLJIT_SP) | IMM(-local_size), DR(SLJIT_SP))); in sljit_emit_enter()
615 offs = local_size - (sljit_sw)sizeof(sljit_sw); in sljit_emit_enter()
[all …]
|
| D | sljitNativeTILEGX_64.c | 1177 sljit_s32 fscratches, sljit_s32 fsaveds, sljit_s32 local_size) in sljit_emit_enter() argument
1183 …eck_sljit_emit_enter(compiler, options, args, scratches, saveds, fscratches, fsaveds, local_size)); in sljit_emit_enter()
1184 set_emit_enter(compiler, options, args, scratches, saveds, fscratches, fsaveds, local_size); in sljit_emit_enter()
1186 local_size += GET_SAVED_REGISTERS_SIZE(scratches, saveds, 1); in sljit_emit_enter()
1187 local_size = (local_size + 7) & ~7; in sljit_emit_enter()
1188 compiler->local_size = local_size; in sljit_emit_enter()
1190 if (local_size <= SIMM_16BIT_MAX) { in sljit_emit_enter()
1192 FAIL_IF(ADDLI(SLJIT_LOCALS_REG_mapped, SLJIT_LOCALS_REG_mapped, -local_size)); in sljit_emit_enter()
1195 FAIL_IF(load_immediate(compiler, TMP_REG1_mapped, local_size)); in sljit_emit_enter()
1199 local_size = 0; in sljit_emit_enter()
[all …]
|
| D | sljitNativePPC_common.c | 610 sljit_s32 fscratches, sljit_s32 fsaveds, sljit_s32 local_size) in sljit_emit_enter() argument
615 …ljit_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size)); in sljit_emit_enter()
616 set_emit_enter(compiler, options, arg_types, scratches, saveds, fscratches, fsaveds, local_size); in sljit_emit_enter()
652 local_size += GET_SAVED_REGISTERS_SIZE(scratches, saveds, 1) + SLJIT_LOCALS_OFFSET; in sljit_emit_enter()
653 local_size = (local_size + 15) & ~0xf; in sljit_emit_enter()
654 compiler->local_size = local_size; in sljit_emit_enter()
657 if (local_size <= SIMM_MAX) in sljit_emit_enter()
658 FAIL_IF(push_inst(compiler, STWU | S(SLJIT_SP) | A(SLJIT_SP) | IMM(-local_size))); in sljit_emit_enter()
660 FAIL_IF(load_immediate(compiler, 0, -local_size)); in sljit_emit_enter()
664 if (local_size <= SIMM_MAX) in sljit_emit_enter()
[all …]
|
| /external/mesa3d/src/gallium/drivers/freedreno/a5xx/ |
| D | fd5_compute.c | 170 const unsigned *local_size = info->block; // v->shader->nir->info->cs.local_size; in fd5_launch_grid() local
176 A5XX_HLSQ_CS_NDRANGE_0_LOCALSIZEX(local_size[0] - 1) | in fd5_launch_grid()
177 A5XX_HLSQ_CS_NDRANGE_0_LOCALSIZEY(local_size[1] - 1) | in fd5_launch_grid()
178 A5XX_HLSQ_CS_NDRANGE_0_LOCALSIZEZ(local_size[2] - 1)); in fd5_launch_grid()
179 OUT_RING(ring, A5XX_HLSQ_CS_NDRANGE_1_SIZE_X(local_size[0] * num_groups[0])); in fd5_launch_grid()
181 OUT_RING(ring, A5XX_HLSQ_CS_NDRANGE_3_SIZE_Y(local_size[1] * num_groups[1])); in fd5_launch_grid()
183 OUT_RING(ring, A5XX_HLSQ_CS_NDRANGE_5_SIZE_Z(local_size[2] * num_groups[2])); in fd5_launch_grid()
199 OUT_RING(ring, A5XX_CP_EXEC_CS_INDIRECT_3_LOCALSIZEX(local_size[0] - 1) | in fd5_launch_grid()
200 A5XX_CP_EXEC_CS_INDIRECT_3_LOCALSIZEY(local_size[1] - 1) | in fd5_launch_grid()
201 A5XX_CP_EXEC_CS_INDIRECT_3_LOCALSIZEZ(local_size[2] - 1)); in fd5_launch_grid()
|
| /external/google-breakpad/src/processor/ |
| D | windows_frame_info.h | 86 local_size(0), in WindowsFrameInfo()
106 local_size(set_local_size), in WindowsFrameInfo()
139 uint32_t local_size = strtoul(tokens[7], NULL, 16); in ParseFromString() local
156 local_size, in ParseFromString()
170 local_size = that.local_size; in CopyFrom()
197 uint32_t local_size; member
|
| /external/deqp/external/openglcts/modules/gles31/ |
| D | es31cComputeShaderTests.cpp | 752 std::string GenSource(const uvec3& local_size, const uvec3& num_groups) in GenSource() argument
754 const uvec3 global_size = local_size * num_groups; in GenSource()
756 ss << NL "layout(local_size_x = " << local_size.x() << ", local_size_y = " << local_size.y() in GenSource()
757 …<< ", local_size_z = " << local_size.z() << ") in;" NL "const uvec3 kGlobalSize = uvec3(" << globa… in GenSource()
792 bool RunIteration(const uvec3& local_size, const uvec3& num_groups, bool dispatch_indirect) in RunIteration() argument
796 m_program = CreateComputeProgram(GenSource(local_size, num_groups)); in RunIteration()
810 …local_size.x() * num_groups.x() * local_size.y() * num_groups.y() * local_size.z() * num_groups.z(… in RunIteration()
853 for (GLuint z = 0; z < local_size.z() * num_groups.z(); ++z) in RunIteration()
855 for (GLuint y = 0; y < local_size.y() * num_groups.y(); ++y) in RunIteration()
857 for (GLuint x = 0; x < local_size.x() * num_groups.x(); ++x) in RunIteration()
[all …]
|
| /external/deqp/external/openglcts/modules/gl/ |
| D | gl4cComputeShaderTests.cpp | 744 std::string GenSource(const uvec3& local_size, const uvec3& num_groups) in GenSource() argument
746 const uvec3 global_size = local_size * num_groups; in GenSource()
748 ss << NL "layout(local_size_x = " << local_size.x() << ", local_size_y = " << local_size.y() in GenSource()
749 …<< ", local_size_z = " << local_size.z() << ") in;" NL "const uvec3 kGlobalSize = uvec3(" << globa… in GenSource()
784 bool RunIteration(const uvec3& local_size, const uvec3& num_groups, bool dispatch_indirect) in RunIteration() argument
788 m_program = CreateComputeProgram(GenSource(local_size, num_groups)); in RunIteration()
811 …local_size.x() * num_groups.x() * local_size.y() * num_groups.y() * local_size.z() * num_groups.z(… in RunIteration()
852 for (GLuint z = 0; z < local_size.z() * num_groups.z(); ++z) in RunIteration()
854 for (GLuint y = 0; y < local_size.y() * num_groups.y(); ++y) in RunIteration()
856 for (GLuint x = 0; x < local_size.x() * num_groups.x(); ++x) in RunIteration()
[all …]
|
| /external/mesa3d/src/amd/vulkan/ |
| D | radv_meta_buffer.c | 20 b.shader->info.cs.local_size[0] = 64; in build_buffer_fill_shader()
21 b.shader->info.cs.local_size[1] = 1; in build_buffer_fill_shader()
22 b.shader->info.cs.local_size[2] = 1; in build_buffer_fill_shader()
27 b.shader->info.cs.local_size[0], in build_buffer_fill_shader()
28 b.shader->info.cs.local_size[1], in build_buffer_fill_shader()
29 b.shader->info.cs.local_size[2], 0); in build_buffer_fill_shader()
72 b.shader->info.cs.local_size[0] = 64; in build_buffer_copy_shader()
73 b.shader->info.cs.local_size[1] = 1; in build_buffer_copy_shader()
74 b.shader->info.cs.local_size[2] = 1; in build_buffer_copy_shader()
79 b.shader->info.cs.local_size[0], in build_buffer_copy_shader()
[all …]
|
| D | radv_meta_bufimage.c | 50 b.shader->info.cs.local_size[0] = 16; in build_nir_itob_compute_shader()
51 b.shader->info.cs.local_size[1] = 16; in build_nir_itob_compute_shader()
52 b.shader->info.cs.local_size[2] = 1; in build_nir_itob_compute_shader()
66 b.shader->info.cs.local_size[0], in build_nir_itob_compute_shader()
67 b.shader->info.cs.local_size[1], in build_nir_itob_compute_shader()
68 b.shader->info.cs.local_size[2], 0); in build_nir_itob_compute_shader()
276 b.shader->info.cs.local_size[0] = 16; in build_nir_btoi_compute_shader()
277 b.shader->info.cs.local_size[1] = 16; in build_nir_btoi_compute_shader()
278 b.shader->info.cs.local_size[2] = 1; in build_nir_btoi_compute_shader()
292 b.shader->info.cs.local_size[0], in build_nir_btoi_compute_shader()
[all …]
|
| D | radv_shader.c | 727 unsigned *local_size = variant->nir->info.cs.local_size; in radv_GetShaderInfoAMD() local
728 unsigned workgroup_size = local_size[0] * local_size[1] * local_size[2]; in radv_GetShaderInfoAMD()
733 statistics.computeWorkGroupSize[0] = local_size[0]; in radv_GetShaderInfoAMD()
734 statistics.computeWorkGroupSize[1] = local_size[1]; in radv_GetShaderInfoAMD()
735 statistics.computeWorkGroupSize[2] = local_size[2]; in radv_GetShaderInfoAMD()
|
| D | radv_meta_resolve_cs.c | 89 b.shader->info.cs.local_size[0] = 16; in build_resolve_compute_shader()
90 b.shader->info.cs.local_size[1] = 16; in build_resolve_compute_shader()
91 b.shader->info.cs.local_size[2] = 1; in build_resolve_compute_shader()
105 b.shader->info.cs.local_size[0], in build_resolve_compute_shader()
106 b.shader->info.cs.local_size[1], in build_resolve_compute_shader()
107 b.shader->info.cs.local_size[2], 0); in build_resolve_compute_shader()
|
| /external/mesa3d/src/intel/compiler/ |
| D | brw_nir_lower_cs_intrinsics.c | 87 unsigned *size = nir->info.cs.local_size; in lower_cs_intrinsics_convert_block()
143 state.local_workgroup_size = nir->info.cs.local_size[0] * in brw_nir_lower_cs_intrinsics()
144 nir->info.cs.local_size[1] * in brw_nir_lower_cs_intrinsics()
145 nir->info.cs.local_size[2]; in brw_nir_lower_cs_intrinsics()
|
| /external/deqp-deps/glslang/Test/baseResults/ |
| D | 320.comp.out | 3 local_size = (1, 1, 1)
14 local_size = (1, 1, 1)
|
| D | negativeArraySize.comp.out | 7 local_size = (1, 1, 1)
18 local_size = (1, 1, 1)
|
| D | 450.comp.out | 8 local_size = (1, 1, 1)
19 local_size = (1, 1, 1)
|
| D | 430.comp.out | 2 ERROR: 0:4: 'local_size' : cannot change previously set size
3 ERROR: 0:5: 'local_size' : too large; see gl_MaxComputeWorkGroupSize
11 ERROR: 0:51: 'local_size' : can only apply to 'in'
12 ERROR: 0:51: 'local_size' : can only apply to 'in'
13 ERROR: 0:51: 'local_size' : can only apply to 'in'
22 local_size = (2, 1, 4096)
154 local_size = (2, 1, 4096)
|
| D | hlsl.localStructuredBuffer.comp.out | 7 local_size = (1, 1, 1)
23 local_size = (1, 1, 1)
|
| /external/mesa3d/src/compiler/glsl/ |
| D | ast_type.cpp | 421 if (q.flags.q.local_size & (1 << i)) { in merge_qualifier()
422 if (this->local_size[i] in merge_qualifier()
424 this->local_size[i]->merge_qualifier(q.local_size[i]); in merge_qualifier()
426 this->local_size[i] = q.local_size[i]; in merge_qualifier()
641 valid_in_mask.flags.q.local_size = 7; in validate_in_qualifier()
714 if (state->in_qualifier->flags.q.local_size) { in merge_into_in_qualifier()
716 state->in_qualifier->local_size); in merge_into_in_qualifier()
717 state->in_qualifier->flags.q.local_size = 0; in merge_into_in_qualifier()
719 state->in_qualifier->local_size[i] = NULL; in merge_into_in_qualifier()
813 bad.flags.q.local_size ? " local_size" : "", in validate_flags()
|
| /external/mesa3d/src/mapi/glapi/gen/ |
| D | gl_x86-64_asm.py | 42 def local_size(registers): function
56 adjust_stack = local_size(registers)
67 adjust_stack = local_size(registers)
|
| /external/mesa3d/src/mesa/drivers/dri/i965/ |
| D | brw_compute.c | 124 unsigned group_size = prog_data->local_size[0] * in brw_emit_gpgpu_walker()
125 prog_data->local_size[1] * prog_data->local_size[2]; in brw_emit_gpgpu_walker()
|