@@ -45,11 +45,20 @@ extern int g_ggml_sycl_debug;
4545extern int g_ggml_sycl_disable_optimize;
4646extern int g_ggml_sycl_prioritize_dmmv;
4747
48- #define GGML_SYCL_DEBUG (...) \
49- do { \
50- if (g_ggml_sycl_debug) \
51- fprintf (stderr, __VA_ARGS__); \
52- } while (0 )
48+ #if defined(__clang__) && __has_builtin(__builtin_expect)
49+ // Hint the optimizer to pipeline the more likely following instruction in branches
50+ #define LIKELY (expr ) __builtin_expect(expr, true )
51+ #define UNLIKELY (expr ) __builtin_expect(expr, false )
52+ #else
53+ #define LIKELY (expr ) (expr)
54+ #define UNLIKELY (expr ) (expr)
55+ #endif
56+
57+ #define GGML_SYCL_DEBUG (...) \
58+ do { \
59+ if (UNLIKELY (g_ggml_sycl_debug)) \
60+ fprintf (stderr, __VA_ARGS__); \
61+ } while (0 )
5362
5463#define CHECK_TRY_ERROR (expr ) \
5564 [&]() { \
@@ -492,8 +501,10 @@ constexpr size_t ceil_div(const size_t m, const size_t n) {
492501
493502bool gpu_has_xmx (sycl::device &dev);
494503
495- template <int N, class T >
496- void debug_print_array (const std::string& prefix, const T array[N]) {
504+ template <int N, class T > void debug_print_array (const std::string & prefix, const T array[N]) {
505+ if (LIKELY (!g_ggml_sycl_debug)) {
506+ return ;
507+ }
497508 std::stringstream ss;
498509 ss << prefix << " =[" ;
499510 for (std::size_t i = 0 ; i < N - 1 ; ++i) {
@@ -506,7 +517,11 @@ void debug_print_array(const std::string& prefix, const T array[N]) {
506517 GGML_SYCL_DEBUG (" %s" , ss.str ().c_str ());
507518}
508519
509- inline void debug_print_tensor (const std::string& prefix, const ggml_tensor* tensor, const std::string& suffix = " " ) {
520+ inline void debug_print_tensor (const std::string & prefix, const ggml_tensor * tensor,
521+ const std::string & suffix = " " ) {
522+ if (LIKELY (!g_ggml_sycl_debug)) {
523+ return ;
524+ }
510525 GGML_SYCL_DEBUG (" %s=" , prefix.c_str ());
511526 if (tensor) {
512527 GGML_SYCL_DEBUG (" '%s':type=%s" , tensor->name , ggml_type_name (tensor->type ));
@@ -526,7 +541,7 @@ inline void debug_print_tensor(const std::string& prefix, const ggml_tensor* ten
526541
527542struct scope_op_debug_print {
528543 scope_op_debug_print (const std::string& func, const ggml_tensor* dst, std::size_t num_src, const std::string& suffix = " " ) : func(func) {
529- if (!g_ggml_sycl_debug) {
544+ if (LIKELY ( !g_ggml_sycl_debug) ) {
530545 return ;
531546 }
532547 GGML_SYCL_DEBUG (" [SYCL][OP] call %s:" , func.c_str ());
0 commit comments