Home
last modified time | relevance | path

Searched refs:skc_uint (Results 1 – 25 of 51) sorted by relevance

123

/external/skqp/src/compute/skc/platforms/cl_12/
Dconfig_cl.h25 skc_uint size;
26 skc_uint subbufs;
29 skc_uint size;
30 skc_uint subbufs;
35 skc_uint size;
39 skc_uint bytes; // bytes per subblock -- pow2
40 skc_uint words; // words per subblock -- pow2
45 skc_uint bytes; // bytes per block -- pow2
46 skc_uint words; // words per block -- pow2
47 skc_uint subblocks; // subblocks per block -- block.bytes >= subblock.bytes
[all …]
Draster_builder_cl_12.h55 skc_uint nodeword;
57 skc_uint tcc;
59 skc_uint transform : SKC_CMD_RASTERIZE_BITS_TRANSFORM;
60 skc_uint clip : SKC_CMD_RASTERIZE_BITS_CLIP;
61 skc_uint cohort : SKC_CMD_RASTERIZE_BITS_COHORT;
95 skc_uint blocks; // # of rk blocks
96 skc_uint offset; // start of rk span
97 skc_uint pk; // # of pk keys
98 skc_uint rk; // # of rk keys
107 skc_uint blocks; // # of blocks in raster -- initially just rk blocks
[all …]
Dhandle_pool_cl_12.h74 skc_uint block;
75 skc_uint rem;
94 skc_uint rem; // # of available records
95 skc_uint head; // index of first record
100 skc_uint index; // index of this record -- never modified
102 skc_uint next; // index of next record
103 skc_uint block; // block index of reclaimed handles
143 skc_uint count;
147 skc_uint * indices; // stack of indices to fixed-size blocks of host handles
148 skc_uint count; // number of handles -- valid from [0,size)
[all …]
Dhandle_pool_cl_12.c113 skc_uint const size, in skc_handle_pool_create()
114 skc_uint const width, in skc_handle_pool_create()
115 skc_uint const recs) in skc_handle_pool_create()
117 skc_uint const blocks = (size + width - 1) / width; in skc_handle_pool_create()
118 skc_uint const blocks_padded = blocks + SKC_HANDLE_POOL_BLOCKS_PAD; in skc_handle_pool_create()
119 skc_uint const handles = blocks * width; in skc_handle_pool_create()
120 skc_uint const handles_padded = blocks_padded * width; in skc_handle_pool_create()
121 skc_uint const recs_padded = recs + 2; // one for pointer and one for head node in skc_handle_pool_create()
131 for (skc_uint ii=0; ii<handles; ii++) in skc_handle_pool_create()
134 for (skc_uint ii=0; ii<handles; ii++) in skc_handle_pool_create()
[all …]
Dblock_pool_cl.h27 skc_uint pool_size; // number of blocks
28 skc_uint ring_pow2; // rounded-up pow2 of pool_size
29 skc_uint ring_mask; // ring_pow2 - 1
41 skc_uint u32a2[2];
44 skc_uint reads;
45 skc_uint writes;
Dpath_builder_cl_12.c134 typedef skc_uint skc_ringdex_t;
142 skc_uint subbuf;
143 skc_uint block;
145 skc_uint block;
146 skc_uint subbuf;
161 skc_uint from; // inclusive starting index : [from,to)
162 skc_uint to; // non-inclusive ending index : [from,to)
187 skc_uint subbufs; // how many subbufs in the buffer?
190 skc_uint buffer; // how many blocks in the buffer?
191 skc_uint subbuf; // how many blocks in a subbuf?
[all …]
Dextent_cl_12.c89 skc_uint const zero = 0; in skc_extent_tdrw_zero()
147 skc_uint const zero = 0; in skc_extent_phr_pdrw_zero()
211 skc_uint const zero = 0; in skc_extent_thr_tdrw_zero()
258 skc_uint const count = skc_extent_ring_snap_count(snap->snap); in skc_extent_phw1g_tdrNs_snap_alloc()
269 skc_uint const index_lo = snap->snap->reads & ring->size.mask; in skc_extent_phw1g_tdrNs_snap_alloc()
270 skc_uint const count_max = ring->size.pow2 - index_lo; in skc_extent_phw1g_tdrNs_snap_alloc()
271 skc_uint const count_lo = min(count_max,count); in skc_extent_phw1g_tdrNs_snap_alloc()
276 skc_uint const bytes_hi = (count - count_max) * ring->size.elem; in skc_extent_phw1g_tdrNs_snap_alloc()
343 skc_uint const count = skc_extent_ring_snap_count(snap->snap); in skc_extent_phrwg_tdrNs_snap_alloc()
354 skc_uint const index_lo = snap->snap->reads & ring->size.mask; in skc_extent_phrwg_tdrNs_snap_alloc()
[all …]
/external/skqp/src/compute/skc/
Dextent_ring.h42 skc_uint reads; // number of reads
43 skc_uint writes; // number of writes
50 skc_uint reads; // number of reads
51 skc_uint writes; // number of writes
56 skc_uint pow2; // ring size must be pow2
57 skc_uint mask; // modulo is a mask because size is pow2
58 skc_uint snap; // max elements in a snapshot (not req'd to be pow2)
59 skc_uint elem; // size of element in bytes
69 skc_uint const size_pow2,
70 skc_uint const size_snap,
[all …]
Dgrid.c106 skc_uint id;
118 skc_uint words[SKC_GRID_SIZE_WORDS]; // 0:inactive, 1:active
119 skc_uint count;
123 skc_uint words[SKC_GRID_SIZE_WORDS]; // 0:inactive, 1:active
124 skc_uint count;
140 skc_uint active[SKC_GRID_SIZE_WORDS]; // 1:inactive, 0:active
142 skc_uint count; // number of active ids
220 skc_uint const handle_pool_size) in skc_grid_deps_create()
241 for (skc_uint id=0; id < SKC_GRID_SIZE_IDS; id++) in skc_grid_deps_create()
246 for (skc_uint ii=0; ii < SKC_GRID_SIZE_WORDS-1; ii++) in skc_grid_deps_create()
[all …]
Dstyling_types.h24 typedef skc_uint skc_layer_id;
25 typedef skc_uint skc_group_id;
33 skc_uint u32;
45 skc_uint opcode : 31;
46 skc_uint final : 1;
55 SKC_STATIC_ASSERT(sizeof(union skc_styling_cmd) == sizeof(skc_uint));
66 skc_uint cmds; // starting index of sequence of command words
82 skc_uint depth;
83 skc_uint base;
98 skc_uint lo; // first layer
[all …]
Dpath.h27 skc_uint handle; // host handle
28 skc_uint blocks; // # of S-segment blocks in path
29 skc_uint nodes; // # of S-segment node blocks -- not including header
30 skc_uint prims; // # of path elements: lines, quads, cubics, rat-quads, rat-cubics
68 #define SKC_PATH_HEAD_WORDS_CALC (sizeof(struct skc_path_head) / sizeof(skc_uint))
69 …ATH_HEAD_OFFSET_HANDLE_CALC (SKC_OFFSET_OF(struct skc_path_head,header.handle) / sizeof(skc_uint))
70 …ATH_HEAD_OFFSET_BLOCKS_CALC (SKC_OFFSET_OF(struct skc_path_head,header.blocks) / sizeof(skc_uint))
71 …ATH_HEAD_OFFSET_NODES_CALC (SKC_OFFSET_OF(struct skc_path_head,header.nodes) / sizeof(skc_uint))
72 …ATH_HEAD_OFFSET_PRIMS_CALC (SKC_OFFSET_OF(struct skc_path_head,header.prims) / sizeof(skc_uint))
73 …ATH_HEAD_OFFSET_IDS_CALC (SKC_OFFSET_OF(struct skc_path_head,tag_ids) / sizeof(skc_uint))
Draster.h42 skc_uint na;
75 skc_uint blocks; // # of blocks -- head+node+skb+pkb -- uint2.lo
76 skc_uint na; // unused -- uint2.hi
77 skc_uint nodes; // # of nodes -- not including header -- uint2.lo
78 skc_uint keys; // # of sk+pk keys -- uint2.hi
111 #define SKC_RASTER_HEAD_WORDS_CALC (sizeof(struct skc_raster_head) / sizeof(skc_uint…
113 …_OFFSET_COUNTS_BLOCKS_CALC (SKC_OFFSET_OF(struct skc_raster_head,header.blocks) / sizeof(skc_uint))
114 …_OFFSET_COUNTS_NODES_CALC (SKC_OFFSET_OF(struct skc_raster_head,header.nodes) / sizeof(skc_uint))
115 …_OFFSET_COUNTS_KEYS_CALC (SKC_OFFSET_OF(struct skc_raster_head,header.keys) / sizeof(skc_uint))
117 …ER_HEAD_OFFSET_BOUNDS_CALC (SKC_OFFSET_OF(struct skc_raster_head,bounds) / sizeof(skc_uint))
[all …]
Dcommon.h79 typedef skc_uint skc_path_h; // host path handle
90 skc_uint tcc;
92 skc_uint transform : SKC_CMD_FILL_BITS_TRANSFORM;
93 skc_uint clip : SKC_CMD_FILL_BITS_CLIP;
94 skc_uint cohort : SKC_CMD_FILL_BITS_COHORT;
103 typedef skc_uint skc_raster_h;
111 skc_uint layer_id;
112 skc_uint tx;
113 skc_uint ty;
Dsuballocator.h41 typedef skc_uint skc_subbuf_size_t; // <4GB
56 skc_uint idx; // ids[] index of subbuf in available state
57 skc_uint inuse;
71 skc_uint avail;
72 skc_uint spare;
75 skc_uint align; // required pow2 alignment
76 skc_uint count; // number of subbufs
92 skc_uint const subbufs,
Dstyling.h35 skc_uint size;
36 skc_uint count;
41 skc_uint size;
42 skc_uint count;
47 skc_uint size;
48 skc_uint count;
Dsuballocator.c32 (skc_uint)ss, \
34 (skc_uint)suballocator->total);
41 (skc_uint)ss, \
43 (skc_uint)suballocator->total);
60 skc_uint const subbufs, in skc_suballocator_create()
79 for (skc_uint ii=0; ii<subbufs; ii++) in skc_suballocator_create()
85 suballocator->align = (skc_uint)align; in skc_suballocator_create()
143 skc_uint avail_rem = suballocator->rem.avail; in skc_suballocator_subbuf_alloc()
144 skc_uint spare_rem = suballocator->rem.spare; in skc_suballocator_subbuf_alloc()
146 for (skc_uint avail_idx=0; avail_idx<avail_rem; avail_idx++) in skc_suballocator_subbuf_alloc()
[all …]
Dextent_ring.c24 skc_uint const size_pow2, in skc_extent_ring_init()
25 skc_uint const size_snap, in skc_extent_ring_init()
26 skc_uint const size_elem) in skc_extent_ring_init()
46 skc_uint
58 skc_uint
64 skc_uint
76 skc_uint
184 skc_uint
190 skc_uint
196 skc_uint
Dweakref.c41 skc_uint index : SKC_WEAKREF_INDEX_BITS;
42 skc_uint na_lo : 32 - SKC_WEAKREF_INDEX_BITS;
43 skc_uint na_hi;
74 skc_uint const index) in skc_weakref_init()
86 skc_uint
90 return (skc_uint)*weakref & SKC_WEAKREF_INDEX_MASK; in skc_weakref_index()
Dblock.h187 typedef skc_uint skc_block_id_t;
193 typedef skc_uint skc_tagged_block_id_t;
197 skc_uint u32;
201 skc_uint tag : SKC_TAGGED_BLOCK_ID_BITS_TAG;
202 skc_uint id : SKC_TAGGED_BLOCK_ID_BITS_ID;
/external/skqp/src/compute/skc/platforms/cl_12/kernels/
Dpaths_copy.cl89 skc_uint
105 skc_uint const bp_idx_mask,
106 skc_uint const bp_reads,
107 skc_uint const bp_off)
109 skc_uint const bp_idx = (bp_reads + bp_off) & bp_idx_mask;
120 skc_uint const bp_elems_idx,
122 skc_uint const pb_elems_idx)
124 for (skc_uint ii=0; ii<SKC_PATHS_COPY_ELEMS_PER_BLOCK; ii+=SKC_PATHS_COPY_SUBGROUP_SIZE)
134 (skc_uint)get_global_id(0),pb_elems_idx,
140 (skc_uint)get_global_id(0),pb_elems_idx,
[all …]
Dplace.cl171 skc_uint scratch[SKC_PLACE_SUBGROUP_SIZE]; // will only use SKC_PLACE_SUBGROUP_SIZE
184 // skc_uint span[SKC_PLACE_SMEM_COUNT_TTPK];
260 skc_uint const sk_idx)
262 skc_uint const lo = smem->lo.sk[sk_idx]; // assumes prefix bit is 0
263 skc_uint const hi = smem->hi.sk[sk_idx];
270 …skc_uint const x = (cmd->tx + SKC_BFE(hi,SKC_TTXK_HI_BITS_X,SKC_TTXK_HI_OFFSET_X)) << SKC_TTCK_HI_…
271 …skc_uint const y = (cmd->ty + SKC_BFE(hi,SKC_TTXK_HI_BITS_Y,SKC_TTXK_HI_OFFSET_Y)) << SKC_TTCK_HI_…
282 skc_uint const pk_idx,
283 skc_uint const dx)
285 skc_uint const lo = smem->lo.pk[pk_idx] & SKC_TTXK_LO_MASK_ID_PREFIX; // assumes prefix bit is 1
[all …]
Dpaths_reclaim.cl107 #define SKC_PATHS_RECLAIM_PACKED_COUNT_DECLARE skc_uint
138 __global skc_uint * const bp_elems, // block pool blocks
139 __global skc_uint volatile * const bp_atomics, // read/write atomics
140 … skc_uint const bp_mask, // pow2 modulo mask for block pool ring
145 skc_uint const reclaim_stride = get_num_sub_groups();
147 …skc_uint const reclaim_stride = get_enqueued_num_sub_groups(); // 2.0 supports non-uniform workgro…
149 skc_uint reclaim_idx = get_group_id(0) * reclaim_stride + get_sub_group_id();
170 skc_uint const head_idx = id * SKC_DEVICE_SUBBLOCK_WORDS + get_sub_group_local_id();
174 skc_uint h##I = bp_elems[head_idx + I * SKC_PATHS_RECLAIM_SUBGROUP_SIZE];
181 skc_uint count_blocks, count_nodes;
[all …]
Dprefix.cl105 skc_uint
205 …skc_uint const offset = pb_id * (SKC_DEVICE_SUBBLOCK_WORDS / SKC_TILE_RATIO) + skc_subgrou…
241 skc_uint * const sk_next,
296 skc_ttsk_v_first(skc_ttsk_v_t * const sk_v, skc_uint const sk_next)
359 skc_uint * const sk_next,
360 skc_uint * const rks_next,
403 skc_blocks_replenish(skc_uint * const blocks_next,
404 skc_uint * const blocks_idx,
406 … skc_uint const bp_mask, // pow2 modulo mask for block pool ring
425 skc_blocks_get_next(skc_uint * const blocks_next,
[all …]
Drasters_reclaim.cl134 #define SKC_RASTERS_RECLAIM_PACKED_COUNT_DECLARE skc_uint
165 … __global skc_uint * const bp_elems, // block pool blocks
166 … __global skc_uint volatile * const bp_atomics, // read/write atomics
167 … skc_uint const bp_mask, // pow2 modulo mask for block pool ring
172 skc_uint const reclaim_stride = get_num_sub_groups();
174 …skc_uint const reclaim_stride = get_enqueued_num_sub_groups(); // 2.0 supports non-uniform workgro…
176 skc_uint reclaim_idx = get_group_id(0) * reclaim_stride + get_sub_group_id();
200 …skc_uint const head_id = id * SKC_DEVICE_SUBBLOCK_WORDS + SKC_RASTERS_RECLAIM_STRIDE_H(get_sub_gro…
204 skc_uint h##I = bp_elems[head_id + SKC_RASTERS_RECLAIM_STRIDE_V_LO(I)];
216 skc_uint count_blocks = sub_group_broadcast(h0,0); // SKC_RASTER_HEAD_OFFSET_COUNTS_NODES
[all …]
/external/skqp/src/compute/skc/platforms/cl_12/kernels/devices/gen9/
Dkernel_cl_12.h72 typedef skc_uint skc_paths_copy_elem;
73 typedef skc_uint skc_pb_idx_v;
166 typedef skc_uint skc_bp_elem_t;
178 typedef skc_uint skc_raster_yx_s;
195 typedef skc_uint skc_bp_elem_t;
197 typedef skc_uint skc_ttsk_lo_t;
198 typedef skc_uint skc_ttsk_hi_t;
200 typedef skc_uint skc_ttpk_lo_t;
201 typedef skc_uint skc_ttpk_hi_t;
203 typedef skc_uint skc_ttxk_lo_t;
[all …]

123