ggml-impl.h 23 KB

123456789101112131415161718192021222324252627282930313233343536373839404142434445464748495051525354555657585960616263646566676869707172737475767778798081828384858687888990919293949596979899100101102103104105106107108109110111112113114115116117118119120121122123124125126127128129130131132133134135136137138139140141142143144145146147148149150151152153154155156157158159160161162163164165166167168169170171172173174175176177178179180181182183184185186187188189190191192193194195196197198199200201202203204205206207208209210211212213214215216217218219220221222223224225226227228229230231232233234235236237238239240241242243244245246247248249250251252253254255256257258259260261262263264265266267268269270271272273274275276277278279280281282283284285286287288289290291292293294295296297298299300301302303304305306307308309310311312313314315316317318319320321322323324325326327328329330331332333334335336337338339340341342343344345346347348349350351352353354355356357358359360361362363364365366367368369370371372373374375376377378379380381382383384385386387388389390391392393394395396397398399400401402403404405406407408409410411412413414415416417418419420421422423424425426427428429430431432433434435436437438439440441442443444445446447448449450451452453454455456457458459460461462463464465466467468469470471472473474475476477478479480481482483484485486487488489490491492493494495496497498499500501502503504505506507508509510511512513514515516517518519520521522523524525526527528529530531532533534535536537538539540541542543544545546547548549550551552553554555556557558559560561562563564565566567568569570571572573574575576577578579580581582583584585586587588589590591592593594595596597598599600601602603604605606607608609610611612613614615616617618619620621622623624625626627628629630631632633634635636637638639640641642643644645646647648649650651652653654655656657658659660661662663664665666667668669670671672673674675676677678679680681682683684685686687688689690691692693694695696697698699700701702703704705706707708709710711712713714715716717718719
  1. #pragma once
  2. // GGML internal header
  3. #include "ggml.h"
  4. #include "gguf.h"
  5. #include <assert.h>
  6. #include <math.h>
  7. #include <stdlib.h> // load `stdlib.h` before other headers to work around MinGW bug: https://sourceforge.net/p/mingw-w64/bugs/192/
  8. #include <stdbool.h>
  9. #include <stdint.h>
  10. #include <string.h>
  11. #ifdef __ARM_FEATURE_SVE
  12. #include <arm_sve.h>
  13. #endif // __ARM_FEATURE_SVE
  14. #if defined(__ARM_NEON) && !defined(__CUDACC__) && !defined(__MUSACC__)
  15. // if YCM cannot find <arm_neon.h>, make a symbolic link to it, for example:
  16. //
  17. // $ ln -sfn /Library/Developer/CommandLineTools/usr/lib/clang/13.1.6/include/arm_neon.h ./src/
  18. //
  19. #include <arm_neon.h>
  20. #endif
  21. #ifdef __cplusplus
  22. extern "C" {
  23. #endif
  24. void ggml_print_backtrace(void);
  25. #ifndef MIN
  26. # define MIN(a, b) ((a) < (b) ? (a) : (b))
  27. #endif
  28. #ifndef MAX
  29. # define MAX(a, b) ((a) > (b) ? (a) : (b))
  30. #endif
  31. // required for mmap as gguf only guarantees 32-byte alignment
  32. #define TENSOR_ALIGNMENT 32
  33. // static_assert should be a #define, but if it's not,
  34. // fall back to the _Static_assert C11 keyword.
  35. // if C99 - static_assert is noop
  36. // ref: https://stackoverflow.com/a/53923785/4039976
  37. #ifndef __cplusplus
  38. #ifndef static_assert
  39. #if defined(__STDC_VERSION__) && (__STDC_VERSION__ >= 201100L)
  40. #define static_assert(cond, msg) _Static_assert(cond, msg)
  41. #else
  42. #define static_assert(cond, msg) struct global_scope_noop_trick
  43. #endif
  44. #endif
  45. #endif
  46. static inline int ggml_up32(int n) {
  47. return (n + 31) & ~31;
  48. }
  49. //static inline int ggml_up64(int n) {
  50. // return (n + 63) & ~63;
  51. //}
  52. static inline int ggml_up(int n, int m) {
  53. // assert m is a power of 2
  54. GGML_ASSERT((m & (m - 1)) == 0);
  55. return (n + m - 1) & ~(m - 1);
  56. }
  57. // TODO: move to ggml.h? (won't be able to inline)
  58. static bool ggml_are_same_layout(const struct ggml_tensor * a, const struct ggml_tensor * b) {
  59. if (a->type != b->type) {
  60. return false;
  61. }
  62. for (int i = 0; i < GGML_MAX_DIMS; i++) {
  63. if (a->ne[i] != b->ne[i]) {
  64. return false;
  65. }
  66. if (a->nb[i] != b->nb[i]) {
  67. return false;
  68. }
  69. }
  70. return true;
  71. }
  72. static bool ggml_op_is_empty(enum ggml_op op) {
  73. switch (op) {
  74. case GGML_OP_NONE:
  75. case GGML_OP_RESHAPE:
  76. case GGML_OP_TRANSPOSE:
  77. case GGML_OP_VIEW:
  78. case GGML_OP_PERMUTE:
  79. return true;
  80. default:
  81. return false;
  82. }
  83. }
  84. static inline float ggml_compute_softplus_f32(float input) {
  85. return (input > 20.0f) ? input : logf(1 + expf(input));
  86. }
  87. //
  88. // logging
  89. //
  90. GGML_ATTRIBUTE_FORMAT(2, 3)
  91. GGML_API void ggml_log_internal (enum ggml_log_level level, const char * format, ...);
  92. GGML_API void ggml_log_callback_default(enum ggml_log_level level, const char * text, void * user_data);
  93. #define GGML_LOG(...) ggml_log_internal(GGML_LOG_LEVEL_NONE , __VA_ARGS__)
  94. #define GGML_LOG_INFO(...) ggml_log_internal(GGML_LOG_LEVEL_INFO , __VA_ARGS__)
  95. #define GGML_LOG_WARN(...) ggml_log_internal(GGML_LOG_LEVEL_WARN , __VA_ARGS__)
  96. #define GGML_LOG_ERROR(...) ggml_log_internal(GGML_LOG_LEVEL_ERROR, __VA_ARGS__)
  97. #define GGML_LOG_DEBUG(...) ggml_log_internal(GGML_LOG_LEVEL_DEBUG, __VA_ARGS__)
  98. #define GGML_LOG_CONT(...) ggml_log_internal(GGML_LOG_LEVEL_CONT , __VA_ARGS__)
  99. #define GGML_DEBUG 0
  100. #if (GGML_DEBUG >= 1)
  101. #define GGML_PRINT_DEBUG(...) GGML_LOG_DEBUG(__VA_ARGS__)
  102. #else
  103. #define GGML_PRINT_DEBUG(...)
  104. #endif
  105. #if (GGML_DEBUG >= 5)
  106. #define GGML_PRINT_DEBUG_5(...) GGML_LOG_DEBUG(__VA_ARGS__)
  107. #else
  108. #define GGML_PRINT_DEBUG_5(...)
  109. #endif
  110. #if (GGML_DEBUG >= 10)
  111. #define GGML_PRINT_DEBUG_10(...) GGML_LOG_DEBUG(__VA_ARGS__)
  112. #else
  113. #define GGML_PRINT_DEBUG_10(...)
  114. #endif
  115. // tensor params
  116. static void ggml_set_op_params(struct ggml_tensor * tensor, const void * params, size_t params_size) {
  117. GGML_ASSERT(tensor != NULL); // silence -Warray-bounds warnings
  118. assert(params_size <= GGML_MAX_OP_PARAMS);
  119. memcpy(tensor->op_params, params, params_size);
  120. }
  121. static int32_t ggml_get_op_params_i32(const struct ggml_tensor * tensor, uint32_t i) {
  122. assert(i < GGML_MAX_OP_PARAMS / sizeof(int32_t));
  123. return ((const int32_t *)(tensor->op_params))[i];
  124. }
  125. static float ggml_get_op_params_f32(const struct ggml_tensor * tensor, uint32_t i) {
  126. assert(i < GGML_MAX_OP_PARAMS / sizeof(float));
  127. return ((const float *)(tensor->op_params))[i];
  128. }
  129. static void ggml_set_op_params_i32(struct ggml_tensor * tensor, uint32_t i, int32_t value) {
  130. assert(i < GGML_MAX_OP_PARAMS / sizeof(int32_t));
  131. ((int32_t *)(tensor->op_params))[i] = value;
  132. }
  133. static void ggml_set_op_params_f32(struct ggml_tensor * tensor, uint32_t i, float value) {
  134. assert(i < GGML_MAX_OP_PARAMS / sizeof(float));
  135. ((float *)(tensor->op_params))[i] = value;
  136. }
  137. struct ggml_map_custom1_op_params {
  138. ggml_custom1_op_t fun;
  139. int n_tasks;
  140. void * userdata;
  141. };
  142. struct ggml_map_custom2_op_params {
  143. ggml_custom2_op_t fun;
  144. int n_tasks;
  145. void * userdata;
  146. };
  147. struct ggml_map_custom3_op_params {
  148. ggml_custom3_op_t fun;
  149. int n_tasks;
  150. void * userdata;
  151. };
  152. struct ggml_custom_op_params {
  153. ggml_custom_op_t fun;
  154. int n_tasks;
  155. void * userdata;
  156. };
  157. // bitset
  158. typedef uint32_t ggml_bitset_t;
  159. static_assert(sizeof(ggml_bitset_t) == 4, "bitset_t constants must be updated");
  160. #define BITSET_SHR 5 // log2(sizeof(ggml_bitset_t)*8)
  161. #define BITSET_MASK (sizeof(ggml_bitset_t)*8 - 1)
  162. static size_t ggml_bitset_size(size_t n) {
  163. return (n + BITSET_MASK) >> BITSET_SHR;
  164. }
  165. static inline bool ggml_bitset_get(const ggml_bitset_t * bitset, size_t i) {
  166. return !!(bitset[i >> BITSET_SHR] & (1u << (i & BITSET_MASK)));
  167. }
  168. static inline void ggml_bitset_set(ggml_bitset_t * bitset, size_t i) {
  169. bitset[i >> BITSET_SHR] |= (1u << (i & BITSET_MASK));
  170. }
  171. static inline void ggml_bitset_clear(ggml_bitset_t * bitset, size_t i) {
  172. bitset[i >> BITSET_SHR] &= ~(1u << (i & BITSET_MASK));
  173. }
  174. // hash set
  175. #define GGML_HASHSET_FULL ((size_t)-1)
  176. #define GGML_HASHSET_ALREADY_EXISTS ((size_t)-2)
  177. struct ggml_hash_set {
  178. size_t size;
  179. ggml_bitset_t * used; // whether or not the keys are in use i.e. set
  180. struct ggml_tensor ** keys; // actual tensors in the set, keys[i] is only defined if ggml_bitset_get(used, i)
  181. };
  182. struct ggml_hash_set ggml_hash_set_new(size_t size);
  183. void ggml_hash_set_free(struct ggml_hash_set * hash_set);
  184. // returns the minimum size for a hash set that can hold min_sz elements
  185. size_t ggml_hash_size(size_t min_sz);
  186. // remove all elements from the hash set
  187. void ggml_hash_set_reset(struct ggml_hash_set * hash_set);
  188. // returns true if key is in the hash set
  189. static bool ggml_hash_contains(const struct ggml_hash_set * hash_set, struct ggml_tensor * key);
  190. // returns GGML_HASHSET_FULL if table is full, otherwise the current index of the key or where it should be inserted
  191. static size_t ggml_hash_find(const struct ggml_hash_set * hash_set, const struct ggml_tensor * key);
  192. // returns GGML_HASHSET_ALREADY_EXISTS if key already exists, index otherwise, asserts if table is full
  193. static size_t ggml_hash_insert(struct ggml_hash_set * hash_set, struct ggml_tensor * key);
  194. // return index, asserts if table is full
  195. static size_t ggml_hash_find_or_insert(struct ggml_hash_set * hash_set, struct ggml_tensor * key);
  196. // hash function for ggml_tensor
  197. static inline size_t ggml_hash(const struct ggml_tensor * p) {
  198. // the last 4 bits are always zero due to alignment
  199. return (size_t)(uintptr_t)p >> 4;
  200. }
  201. static size_t ggml_hash_find(const struct ggml_hash_set * hash_set, const struct ggml_tensor * key) {
  202. size_t h = ggml_hash(key) % hash_set->size;
  203. // linear probing
  204. size_t i = h;
  205. while (ggml_bitset_get(hash_set->used, i) && hash_set->keys[i] != key) {
  206. i = (i + 1) % hash_set->size;
  207. if (i == h) {
  208. // visited all hash table entries -> not found
  209. return GGML_HASHSET_FULL;
  210. }
  211. }
  212. return i;
  213. }
  214. static bool ggml_hash_contains(const struct ggml_hash_set * hash_set, struct ggml_tensor * key) {
  215. size_t i = ggml_hash_find(hash_set, key);
  216. return i != GGML_HASHSET_FULL && ggml_bitset_get(hash_set->used, i);
  217. }
  218. static size_t ggml_hash_insert(struct ggml_hash_set * hash_set, struct ggml_tensor * key) {
  219. size_t h = ggml_hash(key) % hash_set->size;
  220. // linear probing
  221. size_t i = h;
  222. do {
  223. if (!ggml_bitset_get(hash_set->used, i)) {
  224. ggml_bitset_set(hash_set->used, i);
  225. hash_set->keys[i] = key;
  226. return i;
  227. }
  228. if (hash_set->keys[i] == key) {
  229. return GGML_HASHSET_ALREADY_EXISTS;
  230. }
  231. i = (i + 1) % hash_set->size;
  232. } while (i != h);
  233. // visited all hash table entries -> not found
  234. GGML_ABORT("fatal error");
  235. }
  236. static size_t ggml_hash_find_or_insert(struct ggml_hash_set * hash_set, struct ggml_tensor * key) {
  237. size_t h = ggml_hash(key) % hash_set->size;
  238. // linear probing
  239. size_t i = h;
  240. do {
  241. if (!ggml_bitset_get(hash_set->used, i)) {
  242. ggml_bitset_set(hash_set->used, i);
  243. hash_set->keys[i] = key;
  244. return i;
  245. }
  246. if (hash_set->keys[i] == key) {
  247. return i;
  248. }
  249. i = (i + 1) % hash_set->size;
  250. } while (i != h);
  251. // visited all hash table entries -> not found
  252. GGML_ABORT("fatal error");
  253. }
  254. // computation graph
  255. enum ggml_cgraph_eval_order {
  256. GGML_CGRAPH_EVAL_ORDER_LEFT_TO_RIGHT = 0,
  257. GGML_CGRAPH_EVAL_ORDER_RIGHT_TO_LEFT,
  258. GGML_CGRAPH_EVAL_ORDER_COUNT
  259. };
  260. struct ggml_cgraph {
  261. int size; // maximum number of nodes/leafs/grads/grad_accs
  262. int n_nodes; // number of nodes currently in use
  263. int n_leafs; // number of leafs currently in use
  264. struct ggml_tensor ** nodes; // tensors with data that can change if the graph is evaluated
  265. struct ggml_tensor ** grads; // the outputs of these tensors are the gradients of the nodes
  266. struct ggml_tensor ** grad_accs; // accumulators for node gradients
  267. struct ggml_tensor ** leafs; // tensors with constant data
  268. int32_t * use_counts;// number of uses of each tensor, indexed by hash table slot
  269. struct ggml_hash_set visited_hash_set;
  270. enum ggml_cgraph_eval_order order;
  271. };
  272. // returns a slice of cgraph with nodes [i0, i1)
  273. // the slice does not have leafs or gradients
  274. // if you need the gradients, get them from the original graph
  275. struct ggml_cgraph ggml_graph_view(struct ggml_cgraph * cgraph, int i0, int i1);
  276. // ggml-alloc.c: true if the operation can reuse memory from its sources
  277. GGML_API bool ggml_op_can_inplace(enum ggml_op op);
  278. // Memory allocation
  279. GGML_API void * ggml_aligned_malloc(size_t size);
  280. GGML_API void ggml_aligned_free(void * ptr, size_t size);
  281. // FP16 <-> FP32
  282. // ref: https://github.com/Maratyszcza/FP16
  283. static inline float fp32_from_bits(uint32_t w) {
  284. union {
  285. uint32_t as_bits;
  286. float as_value;
  287. } fp32;
  288. fp32.as_bits = w;
  289. return fp32.as_value;
  290. }
  291. static inline uint32_t fp32_to_bits(float f) {
  292. union {
  293. float as_value;
  294. uint32_t as_bits;
  295. } fp32;
  296. fp32.as_value = f;
  297. return fp32.as_bits;
  298. }
  299. static inline float ggml_compute_fp16_to_fp32(ggml_fp16_t h) {
  300. const uint32_t w = (uint32_t) h << 16;
  301. const uint32_t sign = w & UINT32_C(0x80000000);
  302. const uint32_t two_w = w + w;
  303. const uint32_t exp_offset = UINT32_C(0xE0) << 23;
  304. #if (defined(__STDC_VERSION__) && (__STDC_VERSION__ >= 199901L) || defined(__GNUC__) && !defined(__STRICT_ANSI__)) && (!defined(__cplusplus) || __cplusplus >= 201703L)
  305. const float exp_scale = 0x1.0p-112f;
  306. #else
  307. const float exp_scale = fp32_from_bits(UINT32_C(0x7800000));
  308. #endif
  309. const float normalized_value = fp32_from_bits((two_w >> 4) + exp_offset) * exp_scale;
  310. const uint32_t magic_mask = UINT32_C(126) << 23;
  311. const float magic_bias = 0.5f;
  312. const float denormalized_value = fp32_from_bits((two_w >> 17) | magic_mask) - magic_bias;
  313. const uint32_t denormalized_cutoff = UINT32_C(1) << 27;
  314. const uint32_t result = sign |
  315. (two_w < denormalized_cutoff ? fp32_to_bits(denormalized_value) : fp32_to_bits(normalized_value));
  316. return fp32_from_bits(result);
  317. }
  318. static inline ggml_fp16_t ggml_compute_fp32_to_fp16(float f) {
  319. #if (defined(__STDC_VERSION__) && (__STDC_VERSION__ >= 199901L) || defined(__GNUC__) && !defined(__STRICT_ANSI__)) && (!defined(__cplusplus) || __cplusplus >= 201703L)
  320. const float scale_to_inf = 0x1.0p+112f;
  321. const float scale_to_zero = 0x1.0p-110f;
  322. #else
  323. const float scale_to_inf = fp32_from_bits(UINT32_C(0x77800000));
  324. const float scale_to_zero = fp32_from_bits(UINT32_C(0x08800000));
  325. #endif
  326. float base = (fabsf(f) * scale_to_inf) * scale_to_zero;
  327. const uint32_t w = fp32_to_bits(f);
  328. const uint32_t shl1_w = w + w;
  329. const uint32_t sign = w & UINT32_C(0x80000000);
  330. uint32_t bias = shl1_w & UINT32_C(0xFF000000);
  331. if (bias < UINT32_C(0x71000000)) {
  332. bias = UINT32_C(0x71000000);
  333. }
  334. base = fp32_from_bits((bias >> 1) + UINT32_C(0x07800000)) + base;
  335. const uint32_t bits = fp32_to_bits(base);
  336. const uint32_t exp_bits = (bits >> 13) & UINT32_C(0x00007C00);
  337. const uint32_t mantissa_bits = bits & UINT32_C(0x00000FFF);
  338. const uint32_t nonsign = exp_bits + mantissa_bits;
  339. return (sign >> 16) | (shl1_w > UINT32_C(0xFF000000) ? UINT16_C(0x7E00) : nonsign);
  340. }
  341. #define GGML_COMPUTE_FP16_TO_FP32(x) ggml_compute_fp16_to_fp32(x)
  342. #define GGML_COMPUTE_FP32_TO_FP16(x) ggml_compute_fp32_to_fp16(x)
  343. #define GGML_FP16_TO_FP32(x) GGML_COMPUTE_FP16_TO_FP32(x)
  344. #define GGML_FP32_TO_FP16(x) GGML_COMPUTE_FP32_TO_FP16(x)
  345. static inline float ggml_e8m0_to_fp32(uint8_t x) {
  346. uint32_t bits; // Stores the raw bit representation of the float
  347. // Handle special case for minimum exponent (denormalized float)
  348. if (x == 0) {
  349. // Bit pattern for 2^(-127):
  350. // - Sign bit: 0 (positive)
  351. // - Exponent: 0 (denormalized number)
  352. // - Mantissa: 0x400000 (0.5 in fractional form)
  353. // Value = 0.5 * 2^(-126) = 2^(-127)
  354. bits = 0x00400000;
  355. }
  356. // note: disabled as we don't need to handle NaNs
  357. //// Handle special case for NaN (all bits set)
  358. //else if (x == 0xFF) {
  359. // // Standard quiet NaN pattern:
  360. // // - Sign bit: 0
  361. // // - Exponent: all 1s (0xFF)
  362. // // - Mantissa: 0x400000 (quiet NaN flag)
  363. // bits = 0x7FC00000;
  364. //}
  365. // Normalized values (most common case)
  366. else {
  367. // Construct normalized float by shifting exponent into position:
  368. // - Exponent field: 8 bits (positions 30-23)
  369. // - Mantissa: 0 (implicit leading 1)
  370. // Value = 2^(x - 127)
  371. bits = (uint32_t) x << 23;
  372. }
  373. float result; // Final float value
  374. // Safely reinterpret bit pattern as float without type-punning issues
  375. memcpy(&result, &bits, sizeof(float));
  376. return result;
  377. }
  378. // Equal to ggml_e8m0_to_fp32/2
  379. // Useful with MXFP4 quantization since the E0M2 values are doubled
  380. static inline float ggml_e8m0_to_fp32_half(uint8_t x) {
  381. uint32_t bits;
  382. // For x < 2: use precomputed denormal patterns
  383. if (x < 2) {
  384. // 0x00200000 = 2^(-128), 0x00400000 = 2^(-127)
  385. bits = 0x00200000 << x;
  386. }
  387. // For x >= 2: normalized exponent adjustment
  388. else {
  389. // 0.5 * 2^(x-127) = 2^(x-128) = normalized with exponent (x-1)
  390. bits = (uint32_t)(x - 1) << 23;
  391. }
  392. // Note: NaNs are not handled here
  393. float result;
  394. memcpy(&result, &bits, sizeof(float));
  395. return result;
  396. }
  397. #define GGML_E8M0_TO_FP32(x) ggml_e8m0_to_fp32(x)
  398. #define GGML_E8M0_TO_FP32_HALF(x) ggml_e8m0_to_fp32_half(x)
  399. /**
  400. * Converts brain16 to float32.
  401. *
  402. * The bfloat16 floating point format has the following structure:
  403. *
  404. * ┌sign
  405. * │
  406. * │ ┌exponent
  407. * │ │
  408. * │ │ ┌mantissa
  409. * │ │ │
  410. * │┌──┴───┐┌─┴───┐
  411. * 0b0000000000000000 brain16
  412. *
  413. * Since bf16 has the same number of exponent bits as a 32bit float,
  414. * encoding and decoding numbers becomes relatively straightforward.
  415. *
  416. * ┌sign
  417. * │
  418. * │ ┌exponent
  419. * │ │
  420. * │ │ ┌mantissa
  421. * │ │ │
  422. * │┌──┴───┐┌─┴───────────────────┐
  423. * 0b00000000000000000000000000000000 IEEE binary32
  424. *
  425. * For comparison, the standard fp16 format has fewer exponent bits.
  426. *
  427. * ┌sign
  428. * │
  429. * │ ┌exponent
  430. * │ │
  431. * │ │ ┌mantissa
  432. * │ │ │
  433. * │┌─┴─┐┌─┴──────┐
  434. * 0b0000000000000000 IEEE binary16
  435. *
  436. * @see IEEE 754-2008
  437. */
  438. static inline float ggml_compute_bf16_to_fp32(ggml_bf16_t h) {
  439. union {
  440. float f;
  441. uint32_t i;
  442. } u;
  443. u.i = (uint32_t)h.bits << 16;
  444. return u.f;
  445. }
  446. /**
  447. * Converts float32 to brain16.
  448. *
  449. * This is binary identical with Google Brain float conversion.
  450. * Floats shall round to nearest even, and NANs shall be quiet.
  451. * Subnormals aren't flushed to zero, except perhaps when used.
  452. * This code should vectorize nicely if using modern compilers.
  453. */
  454. static inline ggml_bf16_t ggml_compute_fp32_to_bf16(float s) {
  455. ggml_bf16_t h;
  456. union {
  457. float f;
  458. uint32_t i;
  459. } u;
  460. u.f = s;
  461. if ((u.i & 0x7fffffff) > 0x7f800000) { /* nan */
  462. h.bits = (u.i >> 16) | 64; /* force to quiet */
  463. return h;
  464. }
  465. h.bits = (u.i + (0x7fff + ((u.i >> 16) & 1))) >> 16;
  466. return h;
  467. }
  468. #define GGML_FP32_TO_BF16(x) ggml_compute_fp32_to_bf16(x)
  469. #define GGML_BF16_TO_FP32(x) ggml_compute_bf16_to_fp32(x)
  470. static inline int32_t ggml_node_get_use_count(const struct ggml_cgraph * cgraph, int node_idx) {
  471. const struct ggml_tensor * node = cgraph->nodes[node_idx];
  472. size_t hash_pos = ggml_hash_find(&cgraph->visited_hash_set, node);
  473. if (!ggml_bitset_get(cgraph->visited_hash_set.used, hash_pos)) {
  474. return 0;
  475. }
  476. return cgraph->use_counts[hash_pos];
  477. }
  478. // return true if the node's results are only used by N other nodes
  479. // and can be fused into their calculations.
  480. static inline bool ggml_node_has_n_uses(const struct ggml_cgraph * cgraph, int node_idx, int32_t n_uses) {
  481. const struct ggml_tensor * node = cgraph->nodes[node_idx];
  482. // check the use count against how many we're replacing
  483. if (ggml_node_get_use_count(cgraph, node_idx) != n_uses) {
  484. return false;
  485. }
  486. // if node is a view, some other node might be using the intermediate result
  487. // via the view source.
  488. if (node->view_src) {
  489. return false;
  490. }
  491. // If the user requested output for the node, can't fuse
  492. if (node->flags & GGML_TENSOR_FLAG_OUTPUT) {
  493. return false;
  494. }
  495. return true;
  496. }
  497. // Returns true if nodes with indices { node_idxs } are the sequence of ggml_ops in ops[]
  498. // and are fusable. Nodes are considered fusable according to this function if:
  499. // - all nodes except the last have only one use and are not views/outputs (see ggml_node_has_N_uses).
  500. // - all nodes except the last are a src of the following node.
  501. // - all nodes are the same shape.
  502. // TODO: Consider allowing GGML_OP_NONE nodes in between
  503. static inline bool ggml_can_fuse_ext(const struct ggml_cgraph * cgraph, const int * node_idxs, const enum ggml_op * ops, int num_ops) {
  504. for (int i = 0; i < num_ops; ++i) {
  505. if (node_idxs[i] >= cgraph->n_nodes) {
  506. return false;
  507. }
  508. struct ggml_tensor * node = cgraph->nodes[node_idxs[i]];
  509. if (node->op != ops[i]) {
  510. return false;
  511. }
  512. if ((node->flags & GGML_TENSOR_FLAG_COMPUTE) == 0) {
  513. return false;
  514. }
  515. if (i < num_ops - 1 && !ggml_node_has_n_uses(cgraph, node_idxs[i], 1)) {
  516. return false;
  517. }
  518. if (i > 0) {
  519. struct ggml_tensor * prev = cgraph->nodes[node_idxs[i - 1]];
  520. if (node->src[0] != prev && node->src[1] != prev) {
  521. return false;
  522. }
  523. if (!ggml_are_same_shape(node, prev)) {
  524. return false;
  525. }
  526. }
  527. }
  528. return true;
  529. }
  530. // same as above, for sequential indices starting at node_idx
  531. static inline bool ggml_can_fuse(const struct ggml_cgraph * cgraph, int node_idx, const enum ggml_op * ops, int num_ops) {
  532. assert(num_ops < 32);
  533. if (node_idx + num_ops > cgraph->n_nodes) {
  534. return false;
  535. }
  536. int idxs[32];
  537. for (int i = 0; i < num_ops; ++i) {
  538. idxs[i] = node_idx + i;
  539. }
  540. return ggml_can_fuse_ext(cgraph, idxs, ops, num_ops);
  541. }
  542. GGML_API bool ggml_can_fuse_subgraph_ext(const struct ggml_cgraph * cgraph,
  543. const int * node_idxs,
  544. int count,
  545. const enum ggml_op * ops,
  546. const int * outputs,
  547. int num_outputs);
  548. // Returns true if the subgraph formed by {node_idxs} can be fused
  549. // checks whethers all nodes which are not part of outputs can be elided
  550. // by checking if their num_uses are confined to the subgraph
  551. static inline bool ggml_can_fuse_subgraph(const struct ggml_cgraph * cgraph,
  552. int node_idx,
  553. int count,
  554. const enum ggml_op * ops,
  555. const int * outputs,
  556. int num_outputs) {
  557. GGML_ASSERT(count < 32);
  558. if (node_idx + count > cgraph->n_nodes) {
  559. return false;
  560. }
  561. int idxs[32];
  562. for (int i = 0; i < count; ++i) {
  563. idxs[i] = node_idx + i;
  564. }
  565. return ggml_can_fuse_subgraph_ext(cgraph, idxs, count, ops, outputs, num_outputs);
  566. }
  567. #ifdef __cplusplus
  568. }
  569. #endif
  570. #ifdef __cplusplus
  571. #include <array>
  572. #include <initializer_list>
  573. #include <vector>
  574. // nicer C++ syntax for ggml_can_fuse
  575. inline bool ggml_can_fuse(const struct ggml_cgraph * cgraph, int node_idx, std::initializer_list<enum ggml_op> ops) {
  576. return ggml_can_fuse(cgraph, node_idx, ops.begin(), (int)ops.size());
  577. }
  578. inline bool ggml_can_fuse_subgraph(const struct ggml_cgraph * cgraph,
  579. int start_idx,
  580. std::initializer_list<enum ggml_op> ops,
  581. std::initializer_list<int> outputs = {}) {
  582. return ggml_can_fuse_subgraph(cgraph, start_idx, ops.size(), ops.begin(), outputs.begin(), outputs.size());
  583. }
  584. // Return true if the edges in the graph match expectations.
  585. inline bool ggml_check_edges(const struct ggml_cgraph * cgraph,
  586. int start_idx,
  587. std::initializer_list<std::array<int, 3>> edges) {
  588. for (const auto & edge : edges) {
  589. int dst_node = edge[0];
  590. int src_idx = edge[1];
  591. int src_node = edge[2];
  592. if (cgraph->nodes[start_idx + dst_node]->src[src_idx] != cgraph->nodes[start_idx + src_node]) {
  593. return false;
  594. }
  595. }
  596. return true;
  597. }
  598. // expose GGUF internals for test code
  599. GGML_API size_t gguf_type_size(enum gguf_type type);
  600. GGML_API struct gguf_context * gguf_init_from_file_impl(FILE * file, struct gguf_init_params params);
  601. GGML_API void gguf_write_to_buf(const struct gguf_context * ctx, std::vector<int8_t> & buf, bool only_meta);
  602. #endif // __cplusplus