debug_log_from_routines should only happen when log_level > 1

lukstafi · lukstafi · commit 41e59a6659cb · 2024-09-09T19:01:04.000+02:00
diff --git a/arrayjit/lib/backend_utils.ml b/arrayjit/lib/backend_utils.ml
@@ -114,7 +114,7 @@ struct
               | _ -> ()));
     fprintf ppf "@,@]";
     is_global
-
+    
   let compile_main ~traced_store ppf llc : unit =
     let open Stdlib.Format in
     let visited = Hash_set.create (module Tn) in
@@ -129,7 +129,7 @@ struct
       | For_loop { index = i; from_; to_; body; trace_it = _ } ->
           fprintf ppf "@[<2>for (int@ %a = %d;@ %a <= %d;@ ++%a) {@ " pp_index i from_ pp_index i
             to_ pp_index i;
-          if Utils.settings.debug_log_from_routines then
+          if Utils.debug_log_from_routines () then
             if B.logs_to_stdout then
               fprintf ppf {|printf(@[<h>"%s%%d: index %a = %%d\n",@] log_id, %a);@ |}
                 !Utils.captured_log_prefix pp_index i pp_index i
@@ -149,7 +149,7 @@ struct
           let loop_debug_f = debug_float tn.prec in
           let num_closing_braces = pp_top_locals ppf llv in
           let num_typ = Ops.cuda_typ_of_prec tn.prec in
-          if Utils.settings.debug_log_from_routines then (
+          if Utils.debug_log_from_routines () then (
             fprintf ppf "@[<2>{@ @[<2>%s new_set_v =@ %a;@]@ " num_typ loop_f llv;
             let v_code, v_idcs = loop_debug_f llv in
             let pp_args =
@@ -186,7 +186,7 @@ struct
             fprintf ppf "@]@ }@,"
           done
       | Comment message ->
-          if Utils.settings.debug_log_from_routines then
+          if Utils.debug_log_from_routines () then
             if B.logs_to_stdout then
               fprintf ppf {|printf(@[<h>"%s%%d: COMMENT: %s\n",@] log_id);@ |}
                 !Utils.captured_log_prefix
@@ -321,7 +321,7 @@ struct
           ("int " ^ Indexing.symbol_ident s.Indexing.static_symbol, Static_idx s))
     in
     let log_file =
-      if Utils.settings.debug_log_from_routines then
+      if Utils.debug_log_from_routines () then
         [
           ((if B.logs_to_stdout then "int log_id" else "const char* log_file_name"), Log_file_name);
         ]
@@ -346,7 +346,7 @@ struct
     (* FIXME: we should also close the file. *)
     if (not (List.is_empty log_file)) && not B.logs_to_stdout then
       fprintf ppf {|FILE* log_file = fopen(log_file_name, "w");@ |};
-    if Utils.settings.debug_log_from_routines && Utils.settings.log_level > 1 then (
+    if Utils.debug_log_from_routines () then (
       fprintf ppf "/* Debug initial parameter state. */@ ";
       List.iter
         ~f:(function
diff --git a/arrayjit/lib/cc_backend.ml b/arrayjit/lib/cc_backend.ml
@@ -282,7 +282,7 @@ let%diagn_sexp link_compiled ~merge_buffer (prior_context : context) (code : pro
     [%log_result name];
     Backend_utils.check_merge_buffer ~merge_buffer ~code_node:code.lowered.merge_node;
     Indexing.apply run_variadic ();
-    if Utils.settings.debug_log_from_routines then (
+    if Utils.debug_log_from_routines () then (
       Utils.log_trace_tree (Stdio.In_channel.read_lines log_file_name);
       Stdlib.Sys.remove log_file_name)
   in
diff --git a/arrayjit/lib/cuda_backend.cudajit.ml b/arrayjit/lib/cuda_backend.cudajit.ml
@@ -328,7 +328,7 @@ let compile ~name bindings ({ Low_level.traced_store; _ } as lowered) =
   let idx_params = Indexing.bound_symbols bindings in
   let b = Buffer.create 4096 in
   let ppf = Stdlib.Format.formatter_of_buffer b in
-  if Utils.settings.debug_log_from_routines then
+  if Utils.debug_log_from_routines () then
     Stdlib.Format.fprintf ppf "@,__device__ int printf (const char * format, ... );@,";
   let is_global = Hash_set.create (module Tn) in
   let params = Syntax.compile_proc ~name ~is_global ppf idx_params lowered in
@@ -425,12 +425,12 @@ let link_proc ~prior_context ~name ~(params : (string * param_source) list) ~glo
     (* Map.iteri global_arrays ~f:(fun ~key ~data:ptr -> if key.Low_level.zero_initialized then
        Cu.memset_d8_async ptr Unsigned.UChar.zero ~length:(Tn.size_in_bytes key.Low_level.tn)); *)
     [%log "launching the kernel"];
-    (if Utils.settings.debug_log_from_routines then
+    (if Utils.debug_log_from_routines () then
        Utils.add_log_processor ~prefix:log_id_prefix @@ fun _output ->
        [%log_block
          context.label;
          Utils.log_trace_tree _output]);
-    (* if Utils.settings.debug_log_from_routines then Cu.ctx_set_limit CU_LIMIT_PRINTF_FIFO_SIZE
+    (* if Utils.debug_log_from_routines () then Cu.ctx_set_limit CU_LIMIT_PRINTF_FIFO_SIZE
        4096; *)
     Cu.launch_kernel func ~grid_dim_x:1 ~block_dim_x:1 ~shared_mem_bytes:0 context.device.stream
       args;
diff --git a/arrayjit/lib/gcc_backend.gccjit.ml b/arrayjit/lib/gcc_backend.gccjit.ml
@@ -627,7 +627,7 @@ let%diagn_sexp compile_proc ~name ~opt_ctx_arrays ctx bindings ~get_ident
   let fkind = Function.Exported in
   let c_str = Type.(get ctx Const_char_ptr) in
   let log_file_name =
-    if Utils.settings.debug_log_from_routines then
+    if Utils.debug_log_from_routines () then
       Some (Param.create ctx c_str "log_file_name", Log_file_name)
     else None
   in
@@ -857,7 +857,7 @@ let%diagn_sexp link_compiled ~merge_buffer (prior_context : context) (code : pro
     [%log_result name];
     Backend_utils.check_merge_buffer ~merge_buffer ~code_node:code.expected_merge_node;
     Indexing.apply run_variadic ();
-    if Utils.settings.debug_log_from_routines then (
+    if Utils.debug_log_from_routines () then (
       Utils.log_trace_tree (Stdio.In_channel.read_lines log_file_name);
       Stdlib.Sys.remove log_file_name)
   in
diff --git a/arrayjit/lib/utils.ml b/arrayjit/lib/utils.ml
@@ -28,6 +28,9 @@ let mref_add_missing mref key ~f =
 type settings = {
   mutable log_level : int;
   mutable debug_log_from_routines : bool;
+      (** If the [debug_log_from_routines] flag is true _and_ the flag [log_level > 1], backends
+          should generate code (e.g. fprintf statements) to log the execution, and arrange for the
+          logs to be emitted via ppx_minidebug. *)
   mutable debug_memory_locations : bool;
   mutable output_debug_files_in_build_directory : bool;
       (** Writes compilation related files in the [build_files] subdirectory of the run directory
@@ -315,6 +318,7 @@ let restore_settings () =
 
 let () = restore_settings ()
 let with_runtime_debug () = settings.output_debug_files_in_build_directory && settings.log_level > 1
+let debug_log_from_routines () = settings.debug_log_from_routines && settings.log_level > 1
 
 let enable_runtime_debug () =
   settings.output_debug_files_in_build_directory <- true;
@@ -505,7 +509,7 @@ let add_log_processor ~prefix process_logs =
     { log_processor_prefix = prefix; process_logs } :: !captured_log_processors
 
 let capture_stdout_logs ?(never_skip = false) arg =
-  if (not never_skip) && not settings.debug_log_from_routines then arg ()
+  if (not never_skip) && not (debug_log_from_routines ()) then arg ()
   else (
     Stdlib.flush Stdlib.stdout;
     let exitp, entrancep = Unix.pipe () and backup = Unix.dup Unix.stdout in
diff --git a/lib/train.ml b/lib/train.ml
@@ -23,7 +23,7 @@ module CDSL = struct
   let virtualize_settings = Arrayjit.Low_level.virtualize_settings
 
   let enable_all_debugs ?(debug_logs = false) ?(hosted_only = true) () =
-    Utils.set_log_level @@ max 1 @@ Utils.settings.log_level;
+    Utils.set_log_level @@ max 2 @@ Utils.settings.log_level;
     Utils.settings.output_debug_files_in_build_directory <- true;
     if hosted_only then virtualize_settings.enable_device_only <- false;
     if debug_logs then Utils.settings.debug_log_from_routines <- true
