1#ifndef ae2fVK_clspv_clkeys_h
2#define ae2fVK_clspv_clkeys_h
4#ifndef ae2fVK_clspv_IS_OPENCL
5#define ae2fVK_clspv_IS_OPENCL 0
10#define ae2f_CL(...) __VA_ARGS__
16#define ae2f_NCL(...) __VA_ARGS__
69#define CLK_ALL_MEM_FENCE (CLK_LOCAL_MEM_FENCE | CLK_GLOBAL_MEM_FENCE)
71#define work_group_reduce_add(x) 0
#define ae2f_structdef(key, name)
__kernel void kTrainAuto(__global void *glob, __local ae2f_float_t *loc, lrlsz_t lr)
loc ae2f_float_t[lsz - 1][Page]: OutStream ae2f_float_t[lsz - 1][Page]: DeltaStream
#define p_weight
sizeof(ae2f_float_t) * pgsz_sqr * llsz
#define ACT(layer_idx, r, x)
#define ACT_DERIV(layer_idx, r, x)
#define p_outstream
sizeof(ae2f_float_t) * lsz * pgsz
__kernel void kPredictStream(__global void *glob, __local ae2f_float_t *loc, const uint32_t lsz)
loc ae2f_float_t[Page]: inp ae2f_float_t[Page]: out
#define p_deltastream
sizeof(ae2f_float_t) * pgsz * llsz
#define p_goal
sizeof(ae2f_float_t) * pgsz
#define p_bias
sizeof(ae2f_float_t) * pgsz * llsz
__kernel void kPredict(__global void *glob, __local ae2f_float_t *loc, const uint32_t lsz)
loc ae2f_float_t[Page]: inp ae2f_float_t[Page]: out
char STATIC_ASSERT_LRLSZEL_SZ[sizeof(lrlszel_t)==(sizeof(uint32_t) > sizeof(ae2f_float_t) ? sizeof(uint32_t) :sizeof(ae2f_float_t)) ? 1 :-1]
#define p_layerszlist
lsz * sizeof(uint32_t)
char STATIC_ASSERT_LRLSZ_SZ[sizeof(lrlsz_t)==sizeof(lrlszel_t) *3 ? 1 :-1]
__kernel void kFollow(__global void *glob, __local ae2f_float_t *loc, lrlsz_t lr)
loc ae2f_float_t[lsz - 1][Page]: OutStream ae2f_float_t[lsz - 1][Page]: DeltaStream
size_t get_local_linear_id()
size_t get_global_linear_id()
size_t get_num_groups(uint dimindx)
#define CLK_ALL_MEM_FENCE
Contains both LOCAL and GLOBAL.
size_t get_global_offset(uint dimindx)
size_t get_group_id(uint dimindx)
size_t get_global_id(uint dimindx)
#define work_group_reduce_add(x)
size_t get_local_id(uint dimindx)
#define ae2fVK_clspv_IS_OPENCL
size_t get_local_size(uint dimindx)
size_t get_global_size(uint dimindx)
size_t get_enqueued_local_size(uint dimindx)
void barrier(cl_mem_fence_flags flags)
#define __ae2f_AnnSlpFollowOneW_imp(inp, delta, weight, learningrate, inp_sz, inp_idx, out_sz, out_idx)
Weights.
#define __ae2f_AnnSlpFollowOneB_imp(r_bias, delta, learningrate_bias)
#define __ae2f_AnnSlpFetchDeltaOne_imp(v_fetchdelta_0, v_fetchdelta_1, out, out_desired, actderiv_opt, lossderiv, retdelta, oidx, osz)
#define _clMlpGetHD1(r_delta_then, i_weight, i_delta, i_iidx, i_isz, i_oidx, i_osz)
#define _clMlpRvrse(v_tmp, r_delta_then, i_oidx, i_iidx, i_isz, i_actderiv_then, i_inp, i_deltaseed)
delta to delta Propagate
#define clMlpGetHD(...)
GetHidDelta Need no structure.
void clMlpGetHD1_t(ae2f_float_t *const r_delta, const ae2f_float_t *const i_weight, const ae2f_float_t *const i_delta, const size_t i_iidx, const size_t i_isz, const size_t i_oidx, const size_t i_osz)
#define _clMlpGetHD(ONE, r_delta_then, i_weight, i_delta, i_iidx, i_isz, i_oidx, i_osz)
#define _clSlpPredict(v_predict, loc, p_inp, p_weight, p_bias, iidx, isz, oidx, osz, ACT)
ae2f_AnnSlpPredictOne_t _clSlpPredict_t