Repository: exo-lang/exo

Issue:
Empty structs differ in C/C++
Empty structs, such as:

```
typedef struct sgemm_Context {
} sgemm_Context;
```

have different sizes in C and C++. This could lead to ABI issues down the line. If we would generate an empty struct, we should either:

1. use `void*` instead, or
2. add an `int_fast32_t dummy;` field

Discussion:
alexreinking: @yamaguchi1024 - I'm requesting your review just so this is in your queue, but it's not ready to land yet.
yamaguchi1024: Great changes!!
Were we able to remove some of the `-fno-` 's in the cmake scripts for Gemmini as a result?
alexreinking: Not yet. That will be my next PR or two

Patch:
diff --git a/apps/x86/sgemm/alex_sgemm.cpp b/apps/x86/sgemm/alex_sgemm.cpp
index 7a7d350..1f42606 100644
--- a/apps/x86/sgemm/alex_sgemm.cpp
+++ b/apps/x86/sgemm/alex_sgemm.cpp
@@ -99,7 +99,7 @@ static void sgemm_accumulate(
 template <int I_REG, int J_REG>
 void sgemm_micro_kernel_staged_M_N(const float *a, const float *b, float *c,
     long ldc, const long N, const long K) {
-  vec_type inner_c[I_REG][J_REG] = {0};
+  vec_type inner_c[I_REG][J_REG] = {{0}};
 
   const auto Nb = N - 16 * (J_REG - 1);
   short mask = (1 << Nb) - 1;
diff --git a/src/exo/LoopIR_compiler.py b/src/exo/LoopIR_compiler.py
index e783cba..76ae9fe 100644
--- a/src/exo/LoopIR_compiler.py
+++ b/src/exo/LoopIR_compiler.py
@@ -1,7 +1,9 @@
-import os
+import functools
 import re
+import textwrap
 from collections import ChainMap
 from collections import defaultdict
+from dataclasses import dataclass
 from pathlib import Path
 
 from .LoopIR import LoopIR, LoopIR_Do
@@ -15,7 +17,7 @@ from .win_analysis import WindowAnalysis
 
 
 def sanitize_str(s):
-    return re.sub(r'\W', '_', s)
+    return re.sub(r"\W", "_", s)
 
 
 # --------------------------------------------------------------------------- #
@@ -24,25 +26,25 @@ def sanitize_str(s):
 CacheDict = lambda: defaultdict(CacheDict)
 
 op_prec = {
-    "or":  10,
+    "or": 10,
     #
     "and": 20,
     #
-    "==":  30,
+    "==": 30,
     #
-    "<":   40,
-    ">":   40,
-    "<=":  40,
-    ">=":  40,
+    "<": 40,
+    ">": 40,
+    "<=": 40,
+    ">=": 40,
     #
-    "+":   50,
-    "-":   50,
+    "+": 50,
+    "-": 50,
     #
-    "*":   60,
-    "/":   60,
-    "%":   60,
+    "*": 60,
+    "/": 60,
+    "%": 60,
     # unary minus
-    "~":   70,
+    "~": 70,
 }
 
 
@@ -79,7 +81,7 @@ def find_all_subprocs(proc_list):
 
         for sp in LoopIR_SubProcs(proc).result():
             if sp in visited:
-                raise ValueError(f'found call cycle involving {sp.name}')
+                raise ValueError(f"found call cycle involving {sp.name}")
             walk(sp, visited | {proc})
 
     for proc in proc_list:
@@ -88,6 +90,7 @@ def find_all_subprocs(proc_list):
     # Reverse for C declaration order.
     return list(reversed(all_procs))
 
+
 class LoopIR_FindMems(LoopIR_Do):
     def __init__(self, proc):
         self._mems = set()
@@ -193,24 +196,42 @@ def find_all_configs(proc_list):
 # --------------------------------------------------------------------------- #
 # --------------------------------------------------------------------------- #
 
-_window_struct_shorthand = {
-    T.f32: 'f32',
-    T.f64: 'f64',
-    T.i8:  'i8',
-    T.i32: 'i32',
-}
+
+@dataclass(frozen=True)
+class WindowStruct:
+    name: str
+    definition: str
+
+
+@functools.cache
+def _window_struct(typename, ctype, n_dims, is_const) -> WindowStruct:
+    const_kwd = "const " if is_const else ""
+    const_suffix = "c" if is_const else ""
+
+    sname = f"exo_win_{n_dims}{typename}{const_suffix}"
+    sdef = (
+        f"struct {sname}{{\n"
+        f"    {const_kwd}{ctype} * const data;\n"
+        f"    const int_fast32_t strides[{n_dims}];\n"
+        f"}};"
+    )
+
+    return WindowStruct(sname, sdef)
 
 
-def window_struct(basetyp, n_dims):
+def window_struct(base_type, n_dims, is_const) -> WindowStruct:
     assert n_dims >= 1
-    sname = f"exo_win_{n_dims}{_window_struct_shorthand[basetyp]}"
 
-    sdef = (f"struct {sname}{{\n"
-            f"    {basetyp.ctype()} *data;\n"
-            f"    int_fast32_t strides[{n_dims}];\n"
-            f"}};")
+    _window_struct_shorthand = {
+        T.f32: "f32",
+        T.f64: "f64",
+        T.i8: "i8",
+        T.i32: "i32",
+    }
 
-    return sname, sdef
+    return _window_struct(
+        _window_struct_shorthand[base_type], base_type.ctype(), n_dims, is_const
+    )
 
 
 # --------------------------------------------------------------------------- #
@@ -227,8 +248,8 @@ def run_compile(proc_list, h_file_name: str):
 
     body = f'#include "{h_file_name}"\n\n{body}'
 
-    header_guard = f'{lib_name}_H'.upper()
-    fwd_decls = f'''
+    header_guard = f"{lib_name}_H".upper()
+    fwd_decls = f"""
 #pragma once
 #ifndef {header_guard}
 #define {header_guard}
@@ -243,25 +264,41 @@ extern "C" {{
 }}
 #endif
 #endif  // {header_guard}
-'''
+"""
 
     return body, fwd_decls
 
 
+_static_helpers = {
+    "exo_floor_div": textwrap.dedent(
+        """
+        static int exo_floor_div(int num, int quot) {
+          int off = (num>=0)? 0 : quot-1;
+          return (num-off)/quot;
+        }
+        """
+    ),
+    "exo_clamp_32to8": textwrap.dedent(
+        """
+        static int8_t exo_clamp_32to8(int32_t x) {{
+          return (x < -128)? -128 : ((x > 127)? 127 : x);
+        }
+        """
+    ),
+}
+
+
 def compile_to_strings(lib_name, proc_list):
     # Get transitive closure of call-graph
     orig_procs = [id(p) for p in proc_list]
 
-    # Determine name for library context struct
-    ctxt_name = f"{lib_name}_Context"
-
     def from_lines(x):
-        return '\n'.join(x)
+        return "\n".join(x)
 
     proc_list = list(sorted(find_all_subprocs(proc_list), key=lambda x: x.name))
 
     # Header contents
-    ctxt_def = _compile_context_struct(find_all_configs(proc_list), ctxt_name)
+    ctxt_name, ctxt_def = _compile_context_struct(find_all_configs(proc_list), lib_name)
     struct_defns = set()
     public_fwd_decls = []
 
@@ -271,6 +308,8 @@ def compile_to_strings(lib_name, proc_list):
     private_fwd_decls = []
     proc_bodies = []
 
+    needed_helpers = set()
+
     # Compile proc bodies
     seen_procs = set()
     for p in proc_list:
@@ -280,14 +319,16 @@ def compile_to_strings(lib_name, proc_list):
 
         # don't compile instruction procedures, but add a comment.
         if p.instr is not None:
-            argstr = ','.join([str(a.name) for a in p.args])
-            proc_bodies.extend([
-                '',
-                '/* relying on the following instruction..."',
-                f"{p.name}({argstr})",
-                p.instr,
-                "*/",
-            ])
+            argstr = ",".join([str(a.name) for a in p.args])
+            proc_bodies.extend(
+                [
+                    "",
+                    '/* relying on the following instruction..."',
+                    f"{p.name}({argstr})",
+                    p.instr,
+                    "*/",
+                ]
+            )
         else:
             is_public_decl = id(p) in orig_procs
 
@@ -297,6 +338,7 @@ def compile_to_strings(lib_name, proc_list):
             comp = Compiler(p, ctxt_name, is_public_decl=is_public_decl)
             d, b = comp.comp_top()
             struct_defns |= comp.struct_defns()
+            needed_helpers |= comp.needed_helpers()
 
             if is_public_decl:
                 public_fwd_decls.append(d)
@@ -306,9 +348,9 @@ def compile_to_strings(lib_name, proc_list):
             proc_bodies.append(b)
 
     # Structs are just blobs of code... still sort them for output stability
-    struct_defns = list(sorted(struct_defns))
+    struct_defns = [x.definition for x in sorted(struct_defns, key=lambda x: x.name)]
 
-    header_contents = f'''
+    header_contents = f"""
 #include <stdint.h>
 #include <stdbool.h>
 
@@ -333,23 +375,17 @@ def compile_to_strings(lib_name, proc_list):
 {from_lines(ctxt_def)}
 {from_lines(struct_defns)}
 {from_lines(public_fwd_decls)}
-'''
-
-    body_contents = f'''
-static int _floor_div(int num, int quot) {{
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}}
+"""
 
-static int8_t _clamp_32to8(int32_t x) {{
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}}
+    helper_code = [_static_helpers[v] for v in needed_helpers]
 
+    body_contents = f"""
+{from_lines(helper_code)}
 {from_lines(memory_code)}
 {from_lines(builtin_code)}
 {from_lines(private_fwd_decls)}
 {from_lines(proc_bodies)}
-'''
+"""
 
     return header_contents, body_contents
 
@@ -369,9 +405,13 @@ def _compile_memories(mems):
     return memory_code
 
 
-def _compile_context_struct(configs, ctxt_name):
-    ctxt_def = [f"typedef struct {ctxt_name} {{ ",
-                f""]
+def _compile_context_struct(configs, lib_name):
+    if not configs:
+        return "void", []
+
+    ctxt_name = f"{lib_name}_Context"
+    ctxt_def = [f"typedef struct {ctxt_name} {{ ", f""]
+
     seen = set()
     for c in sorted(configs, key=lambda x: x.name()):
         name = c.name()
@@ -386,16 +426,17 @@ def _compile_context_struct(configs, ctxt_name):
             ctxt_def += sdef_lines
             ctxt_def += [""]
         else:
-            ctxt_def += [f"// config '{name}' not materialized",
-                         ""]
+            ctxt_def += [f"// config '{name}' not materialized", ""]
+
     ctxt_def += [f"}} {ctxt_name};"]
-    return ctxt_def
+    return ctxt_name, ctxt_def
 
 
 # --------------------------------------------------------------------------- #
 # --------------------------------------------------------------------------- #
 # Loop IR Compiler
 
+
 class Compiler:
     def __init__(self, proc, ctxt_name, *, is_public_decl):
         assert isinstance(proc, LoopIR.proc)
@@ -409,19 +450,20 @@ class Compiler:
         self._tab = ""
         self._lines = []
         self._scalar_refs = set()
-
+        self._needed_helpers = set()
         self.window_defns = set()
-        self.window_cache = CacheDict()
 
-        assert self.proc.name != None, "expected names for compilation"
+        assert self.proc.name is not None, "expected names for compilation"
         name = self.proc.name
         arg_strs = []
         typ_comments = []
 
         # reserve the first "ctxt" argument
-        self.new_varname(Sym('ctxt'), None)
+        self.new_varname(Sym("ctxt"), None)
         arg_strs.append(f"{ctxt_name} *ctxt")
 
+        self.non_const = set(e.buffer for e in proc.eff.writes + proc.eff.reduces)
+
         for a in proc.args:
             mem = a.mem if a.type.is_numeric() else None
             name_arg = self.new_varname(a.name, typ=a.type, mem=mem)
@@ -435,35 +477,37 @@ class Compiler:
                 if a.type.is_real_scalar():
                     self._scalar_refs.add(a.name)
                 if a.type.is_win():
-                    wintyp = self.get_window_type(a.type)
+                    wintyp = self.get_window_type(a)
                     arg_strs.append(f"struct {wintyp} {name_arg}")
                 else:
+                    const_kwd = "const " if a.name not in self.non_const else ""
                     ctyp = a.type.basetype().ctype()
-                    arg_strs.append(f"{ctyp}* {name_arg}")
+                    arg_strs.append(f"{const_kwd}{ctyp}* {name_arg}")
                 mem = f" @{a.mem.name()}" if a.mem else ""
                 comment_str = f"{name_arg} : {a.type} {mem}"
                 typ_comments.append(comment_str)
 
         for pred in proc.preds:
             if not isinstance(pred, LoopIR.Const):
-                self.add_line(f'EXO_ASSUME({self.comp_e(pred)});')
+                self.add_line(f"EXO_ASSUME({self.comp_e(pred)});")
 
         self.comp_stmts(self.proc.body)
 
-        static_kwd = '' if is_public_decl else 'static '
+        static_kwd = "" if is_public_decl else "static "
 
         # Generate headers here?
-        comment = (f"// {name}(\n" +
-                   ',\n'.join(['//     ' + s for s in typ_comments]) +
-                   '\n'
-                   "// )\n")
-        proc_decl = (comment +
-                     f"{static_kwd}void {name}( {', '.join(arg_strs)} );\n")
-        proc_def = (comment +
-                    f"{static_kwd}void {name}( {', '.join(arg_strs)} ) {{\n" +
-                    "\n".join(self._lines) +
-                    "\n"
-                    "}\n")
+        comment = (
+            f"// {name}(\n" + ",\n".join(["//     " + s for s in typ_comments]) + "\n"
+            "// )\n"
+        )
+        proc_decl = comment + f"{static_kwd}void {name}( {', '.join(arg_strs)} );\n"
+        proc_def = (
+            comment
+            + f"{static_kwd}void {name}( {', '.join(arg_strs)} ) {{\n"
+            + "\n".join(self._lines)
+            + "\n"
+            "}\n"
+        )
 
         self.proc_decl = proc_decl
         self.proc_def = proc_def
@@ -482,6 +526,9 @@ class Compiler:
     def struct_defns(self):
         return self.window_defns
 
+    def needed_helpers(self):
+        return self._needed_helpers
+
     def new_varname(self, symbol, typ, mem=None):
         strnm = str(symbol)
         if strnm not in self.names:
@@ -489,7 +536,7 @@ class Compiler:
         else:
             s = self.names[strnm]
             while s in self.names:
-                m = re.match('^(.*)_([0-9]*)$', s)
+                m = re.match(r"^(.*)_([0-9]*)$", s)
                 if not m:
                     s = s + "_1"
                 else:
@@ -511,10 +558,10 @@ class Compiler:
             self.env = self.env.new_child()
             self.names = self.names.new_child()
             self._tab = self._tab + "  "
-        elif only == 'env':
+        elif only == "env":
             self.env = self.env.new_child()
             self.names = self.names.new_child()
-        elif only == 'tab':
+        elif only == "tab":
             self._tab = self._tab + "  "
         else:
             assert False, f"BAD only parameter {only}"
@@ -562,30 +609,27 @@ class Compiler:
         return acc
 
     def get_window_type(self, typ):
+        assert isinstance(typ, T.Window) or (
+            isinstance(typ, LoopIR.fnarg) and typ.type.is_win()
+        )
+
         if isinstance(typ, T.Window):
             base = typ.as_tensor.basetype()
             n_dims = len(typ.as_tensor.shape())
-        elif isinstance(typ, T.Tensor) and typ.is_window:
-            base = typ.basetype()
-            n_dims = len(typ.shape())
+            is_const = typ.src_buf not in self.non_const
         else:
-            assert False, f"not a window type: {typ}"
+            base = typ.type.basetype()
+            n_dims = len(typ.type.shape())
+            is_const = typ.name not in self.non_const
 
-        lookup = self.window_cache[base][n_dims]
-        if isinstance(lookup, str):
-            return lookup
-        else:
-            name, defn = window_struct(base, n_dims)
-            self.window_defns.add(defn)
-            self.window_cache[base][n_dims] = name
-            return name
+        win = window_struct(base, n_dims, is_const)
+        self.window_defns.add(win)
+        return win.name
 
     def comp_s(self, s):
-        styp = type(s)
-
-        if styp is LoopIR.Pass:
+        if isinstance(s, LoopIR.Pass):
             self.add_line("; // NO-OP")
-        elif styp is LoopIR.Assign or styp is LoopIR.Reduce:
+        elif isinstance(s, (LoopIR.Assign, LoopIR.Reduce)):
             if s.name in self._scalar_refs:
                 lhs = f"*{self.env[s.name]}"
             elif self.envtyp[s.name].is_real_scalar():
@@ -602,20 +646,21 @@ class Compiler:
                 assert s.rhs.type.is_real_scalar()
 
                 if lbtyp == T.i8 and rbtyp == T.i32:
-                    rhs = f"_clamp_32to8({rhs})"
+                    rhs = self._call_static_helper("exo_clamp_32to8", rhs)
                 else:
                     rhs = f"({lbtyp.ctype()})({rhs})"
 
             mem: Memory = self.mems[s.name]
-            if styp is LoopIR.Assign:
+            if isinstance(s, LoopIR.Assign):
                 self.add_line(mem.write(s, lhs, rhs))
             else:
                 self.add_line(mem.reduce(s, lhs, rhs))
 
-        elif styp is LoopIR.WriteConfig:
+        elif isinstance(s, LoopIR.WriteConfig):
             if not s.config.is_allow_rw():
-                raise ConfigError(f"{s.srcinfo}: cannot write to config "
-                                  f"'{s.config.name()}'")
+                raise ConfigError(
+                    f"{s.srcinfo}: cannot write to config '{s.config.name()}'"
+                )
 
             nm = s.config.name()
             rhs = self.comp_e(s.rhs)
@@ -628,20 +673,20 @@ class Compiler:
                 assert rtyp.is_real_scalar()
 
                 if ltyp == T.i8 and rtyp == T.i32:
-                    rhs = f"_clamp_32to8({rhs})"
+                    rhs = self._call_static_helper("exo_clamp_32to8", rhs)
                 else:
                     rhs = f"({ltyp.ctype()})({rhs})"
 
             self.add_line(f"ctxt->{nm}.{s.field} = {rhs};")
 
-        elif styp is LoopIR.WindowStmt:
+        elif isinstance(s, LoopIR.WindowStmt):
             win_struct = self.get_window_type(s.rhs.type)
             rhs = self.comp_e(s.rhs)
             assert isinstance(s.rhs, LoopIR.WindowExpr)
             mem = self.mems[s.rhs.name]
             lhs = self.new_varname(s.lhs, typ=s.rhs.type, mem=mem)
             self.add_line(f"struct {win_struct} {lhs} = {rhs};")
-        elif styp is LoopIR.If:
+        elif isinstance(s, LoopIR.If):
             cond = self.comp_e(s.cond)
             self.add_line(f"if ({cond}) {{")
             self.push()
@@ -654,41 +699,36 @@ class Compiler:
                 self.pop()
             self.add_line("}")
 
-        elif styp is LoopIR.Seq:
+        elif isinstance(s, LoopIR.Seq):
             hi = self.comp_e(s.hi)
-            self.push(only='env')
+            self.push(only="env")
             itr = self.new_varname(s.iter, typ=T.index)  # allocate a new string
             self.add_line(f"for (int {itr} = 0; {itr} < {hi}; {itr}++) {{")
-            self.push(only='tab')
+            self.push(only="tab")
             self.comp_stmts(s.body)
             self.pop()
             self.add_line("}")
 
-        elif styp is LoopIR.Alloc:
+        elif isinstance(s, LoopIR.Alloc):
             name = self.new_varname(s.name, typ=s.type, mem=s.mem)
             assert s.type.basetype().is_real_scalar()
             assert s.type.basetype() != T.R
             ctype = s.type.basetype().ctype()
             mem = s.mem or DRAM
-            line = mem.alloc(name,
-                             ctype,
-                             self.shape_strs(s.type.shape()),
-                             s.srcinfo)
+            line = mem.alloc(name, ctype, self.shape_strs(s.type.shape()), s.srcinfo)
 
             self.add_line(line)
-        elif styp is LoopIR.Free:
+        elif isinstance(s, LoopIR.Free):
             name = self.env[s.name]
             assert s.type.basetype().is_real_scalar()
             ctype = s.type.basetype().ctype()
             mem = s.mem or DRAM
-            line = mem.free(name,
-                            ctype,
-                            self.shape_strs(s.type.shape()),
-                            s.srcinfo)
+            line = mem.free(name, ctype, self.shape_strs(s.type.shape()), s.srcinfo)
             self.add_line(line)
-        elif styp is LoopIR.Call:
-            assert all(a.type.is_win() == fna.type.is_win()
-                       for a, fna in zip(s.args, s.f.args))
+        elif isinstance(s, LoopIR.Call):
+            assert all(
+                a.type.is_win() == fna.type.is_win() for a, fna in zip(s.args, s.f.args)
+            )
             args = [self.comp_e(e, call_arg=True) for e in s.args]
             if s.f.instr is not None:
                 d = dict()
@@ -700,10 +740,10 @@ class Compiler:
                     if arg_type.is_win():
                         assert isinstance(s.args[i], LoopIR.WindowExpr)
                         data, _ = self.window_struct_fields(s.args[i])
-                        d[f'{arg_name}_data'] = data
-                        d[f'{arg_name}_int'] = self.env[s.args[i].name]
+                        d[f"{arg_name}_data"] = data
+                        d[f"{arg_name}_int"] = self.env[s.args[i].name]
                     else:
-                        d[f'{arg_name}_data'] = f"({args[i]})"
+                        d[f"{arg_name}_data"] = f"({args[i]})"
 
                 self.add_line(f"{s.f.instr.format(**d)}")
             else:
@@ -740,8 +780,10 @@ class Compiler:
                 mem: Memory = self.mems[e.name]
 
                 if not mem.can_read():
-                    raise MemGenError(f"{e.srcinfo}: cannot read from buffer "
-                                      f"'{e.name}' in memory '{mem.name()}'")
+                    raise MemGenError(
+                        f"{e.srcinfo}: cannot read from buffer "
+                        f"'{e.name}' in memory '{mem.name()}'"
+                    )
 
                 if e.name in self._scalar_refs:
                     return f"*{self.env[e.name]}"
@@ -751,16 +793,15 @@ class Compiler:
                     return self.access_str(e.name, e.idx)
         elif etyp is LoopIR.WindowExpr:
             win_struct = self.get_window_type(e.type)
-            cast = f'({self.envtyp[e.name].basetype().ctype()}*)'
             data, strides = self.window_struct_fields(e)
-            return f"(struct {win_struct}){{ {cast}&{data}, {{ {strides} }} }}"
+            return f"(struct {win_struct}){{ &{data}, {{ {strides} }} }}"
         elif etyp is LoopIR.Const:
             if isinstance(e.val, bool):
-                return 'true' if e.val else 'false'
+                return "true" if e.val else "false"
             return str(e.val)
         elif etyp is LoopIR.BinOp:
             local_prec = op_prec[e.op]
-            int_div = (e.op == "/" and not e.type.is_numeric())
+            int_div = e.op == "/" and not e.type.is_numeric()
             if int_div:
                 local_prec = 0
             op = e.op
@@ -775,8 +816,8 @@ class Compiler:
             if int_div:
                 if isinstance(e.lhs.type, LoopIR.Size):
                     # TODO: too many parens?
-                    return f'(({lhs}) / ({rhs}))'
-                return f"_floor_div({lhs}, {rhs})"
+                    return f"(({lhs}) / ({rhs}))"
+                return self._call_static_helper("exo_floor_div", lhs, rhs)
 
             s = f"{lhs} {op} {rhs}"
             if local_prec < prec:
@@ -797,13 +838,18 @@ class Compiler:
             return strides[e.dim]
         elif etyp is LoopIR.ReadConfig:
             if not e.config.is_allow_rw():
-                raise ConfigError(f"{e.srcinfo}: cannot read from config "
-                                  f"'{e.config.name()}'")
+                raise ConfigError(
+                    f"{e.srcinfo}: cannot read from config '{e.config.name()}'"
+                )
             return f"ctxt->{e.config.name()}.{e.field}"
 
         else:
             assert False, "bad case"
 
+    def _call_static_helper(self, helper, *args):
+        self._needed_helpers.add(helper)
+        return f'{helper}({", ".join(map(str, args))})'
+
     def window_struct_fields(self, e):
         base = self.env[e.name]
         basetyp = self.envtyp[e.name]
@@ -818,6 +864,7 @@ class Compiler:
         all_strides = self.get_strides(base, basetyp, prec=0)
         assert 0 < len(all_strides) == len(e.idx)
         dataptr = mem.window(basetyp, base, idxs, all_strides, e.srcinfo)
-        strides = ', '.join(s for s, w in zip(all_strides, e.idx)
-                            if isinstance(w, LoopIR.Interval))
+        strides = ", ".join(
+            s for s, w in zip(all_strides, e.idx) if isinstance(w, LoopIR.Interval)
+        )
         return dataptr, strides


Test patch:
diff --git a/tests/amx/test_amx_instr.py b/tests/amx/test_amx_instr.py
index 7ab516f..fbe8c1e 100644
--- a/tests/amx/test_amx_instr.py
+++ b/tests/amx/test_amx_instr.py
@@ -29,13 +29,12 @@ def test_ldst_i8_16x64(compiler, sde64):
         st_i8(16, 64, tile1, z)
 
     t = AMXTestBuilder(compiler.basename)
-    t.add_body([f"{compiler.basename}_Context *ctxt;"])
 
     t.alloc_dram_2i8('x', 16, 64, 'i+j')
     t.alloc_dram_2i8('y', 16, 64, '0')
     t.alloc_dram_2i8('z', 16, 64, '0')
 
-    t.add_body(['ldst_i8_16x64(ctxt, x, y, z);',
+    t.add_body(['ldst_i8_16x64(NULL, x, y, z);',
                 '',
                 'if(check_eq_2i8(16, 64, x, z)) {',
                 '    printf("Correct\\n");',
@@ -66,14 +65,13 @@ def test_dpbssd_i8_16x64(compiler, sde64):
         st_i32(16, 16, tile2, z)
 
     t = AMXTestBuilder(compiler.basename)
-    t.add_body([f"{compiler.basename}_Context *ctxt;"])
 
     t.alloc_dram_2i8('x', 16, 64, '1')
     t.alloc_dram_2i8('y', 16, 64, '1')
     t.alloc_dram_2i32('z', 16, 16, '64')  # expected result
     t.alloc_dram_2i32('res', 16, 16, '0')
 
-    t.add_body(['dpbssd_i8_16x64(ctxt, x, y, res);',
+    t.add_body(['dpbssd_i8_16x64(NULL, x, y, res);',
                 '',
                 'if(check_eq_2i32(16, 16, z, res)) {',
                 '    printf("Correct\\n");',
@@ -104,12 +102,11 @@ def test_transform_memory(compiler, sde64):
                     dest[i, 4 * j + k] = src[4 * i + k, j]
 
     t = AMXTestBuilder(compiler.basename)
-    t.add_body([f"{compiler.basename}_Context *ctxt;"])
 
     t.alloc_dram_2i8('src', 64, 16, '5')
     t.alloc_dram_2i8('ans', 16, 64, '5')
     t.alloc_dram_2i8('res', 16, 64, '0')
-    t.add_body(['call_transform_memory(ctxt, 16, 16, src, res);',
+    t.add_body(['call_transform_memory(NULL, 16, 16, src, res);',
                 '',
                 'if(check_eq_2i8(16, 64, ans, res)) {',
                 '    printf("Correct\\n");',
@@ -248,7 +245,6 @@ def test_matmul_on_amx_by_hand_i8(compiler, sde64):
     size2 = 256
 
     t = AMXTestBuilder(compiler.basename)
-    t.add_body([f"{compiler.basename}_Context *ctxt;"])
 
     t.alloc_dram_2i8('x', size1, size2, 'i+j')
     t.alloc_dram_2i8('y_orig', size2, size1, 'j+1')  # before transform_memory
@@ -256,11 +252,11 @@ def test_matmul_on_amx_by_hand_i8(compiler, sde64):
     t.alloc_dram_2i32('z', size1, size1, '0')  # expected result
     t.alloc_dram_2i32('res', size1, size1, '0')
 
-    t.add_body([f'transform_memory(ctxt, {size2}, {size1}, y_orig, y);'])
+    t.add_body([f'transform_memory(NULL, {size2}, {size1}, y_orig, y);'])
     t.add_body(
-        [f'matmul_on_cpu(ctxt, {size1}, {size2}, {size1}, x, y_orig, z);'])
+        [f'matmul_on_cpu(NULL, {size1}, {size2}, {size1}, x, y_orig, z);'])
     t.add_body(
-        [f'matmul_on_amx(ctxt, {size1}, {size2 // 4}, {size1}, x, y, res);'])
+        [f'matmul_on_amx(NULL, {size1}, {size2 // 4}, {size1}, x, y, res);'])
 
     t.add_body([f'if(check_eq_2i32({size1}, {size1}, z, res)) {{',
                 r'    printf("Correct\n");',
@@ -365,7 +361,6 @@ def test_gen_matmul_i8_amx(matmul_i8):
 
 def test_matmul_on_amx_scheduled_i8(compiler, sde64, matmul_i8):
     t = AMXTestBuilder(compiler.basename)
-    t.add_body([f"{compiler.basename}_Context *ctxt;"])
 
     # duplicate from the fixture above
     size1 = 256
@@ -377,11 +372,11 @@ def test_matmul_on_amx_scheduled_i8(compiler, sde64, matmul_i8):
     t.alloc_dram_2i32('z', size1, size1, '0')  # expected result
     t.alloc_dram_2i32('res', size1, size1, '0')
 
-    t.add_body([f'transform_memory(ctxt, {size2}, {size1}, y_orig, y);'])
+    t.add_body([f'transform_memory(NULL, {size2}, {size1}, y_orig, y);'])
     t.add_body(
-        [f'matmul_on_cpu(ctxt, {size1}, {size2}, {size1}, x, y_orig, z);'])
+        [f'matmul_on_cpu(NULL, {size1}, {size2}, {size1}, x, y_orig, z);'])
     t.add_body(
-        [f'matmul_on_amx(ctxt, x, y, res);'])
+        [f'matmul_on_amx(NULL, x, y, res);'])
 
     t.add_body([f'if(check_eq_2i32({size1}, {size1}, z, res)) {{',
                 '    printf("Correct\\n");',
diff --git a/tests/golden/pldi22/test_gemmini_conv_ae/test_conv_ae.txt b/tests/golden/pldi22/test_gemmini_conv_ae/test_conv_ae.txt
index 9c60de2..3777d24 100644
--- a/tests/golden/pldi22/test_gemmini_conv_ae/test_conv_ae.txt
+++ b/tests/golden/pldi22/test_gemmini_conv_ae/test_conv_ae.txt
@@ -41,17 +41,21 @@ typedef struct c_code_str_Context {
     } ConfigStore;
 
 } c_code_str_Context;
-struct exo_win_2i32{
-    int32_t *data;
-    int_fast32_t strides[2];
+struct exo_win_2i32c{
+    const int32_t * const data;
+    const int_fast32_t strides[2];
 };
 struct exo_win_2i8{
-    int8_t *data;
-    int_fast32_t strides[2];
+    int8_t * const data;
+    const int_fast32_t strides[2];
 };
-struct exo_win_3i8{
-    int8_t *data;
-    int_fast32_t strides[3];
+struct exo_win_2i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_3i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[3];
 };
 // conv_on_gemmini(
 //     output : i8[4,56,56,64]  @DRAM,
@@ -61,18 +65,10 @@ struct exo_win_3i8{
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_on_gemmini( c_code_str_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale );
-
+void conv_on_gemmini( c_code_str_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale );
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
 
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -119,7 +115,7 @@ gemmini_extended_config_st({dst_stride}, {act}, {scale}[0]);
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_on_gemmini( c_code_str_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale ) {
+void conv_on_gemmini( c_code_str_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale ) {
 gemmini_extended_config_st((64), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -156,7 +152,7 @@ for (int b = 0; b < 4; b++) {
             }
           }
           if (orow_ii + 7 * orow_io == 0 || false) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((0) * (3 * 4 * 4 * 16 * 16) + (0) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -204,7 +200,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || false) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (1) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((0) * (3 * 4 * 4 * 16 * 16) + (1) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -252,7 +248,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || false) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (2) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((0) * (3 * 4 * 4 * 16 * 16) + (2) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -300,7 +296,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || false) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((1 + orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((1) * (3 * 4 * 4 * 16 * 16) + (0) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -348,7 +344,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || false) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((1 + orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (1) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((1) * (3 * 4 * 4 * 16 * 16) + (1) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -396,7 +392,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || false) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((1 + orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (2) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((1) * (3 * 4 * 4 * 16 * 16) + (2) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -444,7 +440,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || true) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((2 + orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((2) * (3 * 4 * 4 * 16 * 16) + (0) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -492,7 +488,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || true) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (1 + 16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((2 + orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (1) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((2) * (3 * 4 * 4 * 16 * 16) + (1) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -540,7 +536,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_ii + 7 * orow_io == 0 || true) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (2 + 16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((2 + orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (2) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((2) * (3 * 4 * 4 * 16 * 16) + (2) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -601,7 +597,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || false) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((0) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((0) * (3 * 4 * 4 * 16 * 16) + (0) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -647,7 +643,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || false) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((1) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((0) * (3 * 4 * 4 * 16 * 16) + (1) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -693,7 +689,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || false) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((2) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((0) * (3 * 4 * 4 * 16 * 16) + (2) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -739,7 +735,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || false) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((0) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((1) * (3 * 4 * 4 * 16 * 16) + (0) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -785,7 +781,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || false) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((1) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((1) * (3 * 4 * 4 * 16 * 16) + (1) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -831,7 +827,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || false) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (1 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((2) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((1) * (3 * 4 * 4 * 16 * 16) + (2) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -877,7 +873,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || true) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((0) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((2) * (3 * 4 * 4 * 16 * 16) + (0) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -923,7 +919,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || true) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (49) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((1) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((2) * (3 * 4 * 4 * 16 * 16) + (1) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
@@ -969,7 +965,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
           }
         }
         if (orow_ii + 7 * orow_io == 0 || true) {
-          gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+          gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (2 + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (50) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((2) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
         }
         gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s_1)) + ((2) * (3 * 4 * 4 * 16 * 16) + (2) * (4 * 4 * 16 * 16) + (0) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res_1)) + ((0) * (9 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (8));
diff --git a/tests/golden/pldi22/test_gemmini_matmul_ae/test_matmul_ae.txt b/tests/golden/pldi22/test_gemmini_matmul_ae/test_matmul_ae.txt
index 8f53467..95ae99b 100644
--- a/tests/golden/pldi22/test_gemmini_matmul_ae/test_matmul_ae.txt
+++ b/tests/golden/pldi22/test_gemmini_matmul_ae/test_matmul_ae.txt
@@ -45,17 +45,21 @@ typedef struct c_code_str_Context {
     } ConfigStore;
 
 } c_code_str_Context;
-struct exo_win_2i32{
-    int32_t *data;
-    int_fast32_t strides[2];
+struct exo_win_2i32c{
+    const int32_t * const data;
+    const int_fast32_t strides[2];
 };
 struct exo_win_2i8{
-    int8_t *data;
-    int_fast32_t strides[2];
+    int8_t * const data;
+    const int_fast32_t strides[2];
 };
-struct exo_win_3i8{
-    int8_t *data;
-    int_fast32_t strides[3];
+struct exo_win_2i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_3i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[3];
 };
 // matmul_on_gemmini(
 //     scale : f32  @DRAM,
@@ -64,18 +68,10 @@ struct exo_win_3i8{
 //     B : i8[512,512]  @DRAM,
 //     C : i8[512,512]  @DRAM
 // )
-void matmul_on_gemmini( c_code_str_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
-
+void matmul_on_gemmini( c_code_str_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
 
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -153,7 +149,7 @@ gemmini_extended_mvin( 0, ((uint64_t) &{dst_data}),{m}, {n} );
 //     B : i8[512,512]  @DRAM,
 //     C : i8[512,512]  @DRAM
 // )
-void matmul_on_gemmini( c_code_str_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void matmul_on_gemmini( c_code_str_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 gemmini_extended_config_st((512), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
diff --git a/tests/golden/test_apps/test_gemmini_conv.txt b/tests/golden/test_apps/test_gemmini_conv.txt
index 3879ecc..e196dc5 100644
--- a/tests/golden/test_apps/test_gemmini_conv.txt
+++ b/tests/golden/test_apps/test_gemmini_conv.txt
@@ -50,17 +50,21 @@ typedef struct test_case_Context {
     } ConfigStore;
 
 } test_case_Context;
-struct exo_win_2i32{
-    int32_t *data;
-    int_fast32_t strides[2];
+struct exo_win_2i32c{
+    const int32_t * const data;
+    const int_fast32_t strides[2];
 };
 struct exo_win_2i8{
-    int8_t *data;
-    int_fast32_t strides[2];
+    int8_t * const data;
+    const int_fast32_t strides[2];
 };
-struct exo_win_3i8{
-    int8_t *data;
-    int_fast32_t strides[3];
+struct exo_win_2i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_3i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[3];
 };
 // conv_17(
 //     output : i8[4,28,28,128]  @DRAM,
@@ -70,7 +74,7 @@ struct exo_win_3i8{
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_17( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale );
+void conv_17( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale );
 
 // conv_17_cpu(
 //     output : i8[4,28,28,128]  @DRAM,
@@ -80,7 +84,7 @@ void conv_17( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* in
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_17_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale );
+void conv_17_cpu( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale );
 
 // conv_3(
 //     output : i8[4,56,56,64]  @DRAM,
@@ -90,7 +94,7 @@ void conv_17_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_3( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale );
+void conv_3( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale );
 
 // conv_30(
 //     output : i8[4,14,14,256]  @DRAM,
@@ -100,7 +104,7 @@ void conv_3( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_30( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale );
+void conv_30( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale );
 
 // conv_30_cpu(
 //     output : i8[4,14,14,256]  @DRAM,
@@ -110,7 +114,7 @@ void conv_30( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* in
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_30_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale );
+void conv_30_cpu( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale );
 
 // conv_3_cpu(
 //     output : i8[4,56,56,64]  @DRAM,
@@ -120,7 +124,7 @@ void conv_30_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_3_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale );
+void conv_3_cpu( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale );
 
 
 
@@ -132,14 +136,6 @@ void conv_3_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t*
 #include "test_case.h"
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -162,7 +158,7 @@ double _select_(double x, double v, double y, double z) {
 //     src : f32  @DRAM,
 //     dst : i8  @DRAM
 // )
-static void clamp( test_case_Context *ctxt, float* src, int8_t* dst );
+static void clamp( test_case_Context *ctxt, const float* src, int8_t* dst );
 
 
 /* relying on the following instruction..."
@@ -173,7 +169,7 @@ acc_scale(src,dst,scale)
 //     src : f32  @DRAM,
 //     dst : i8  @DRAM
 // )
-static void clamp( test_case_Context *ctxt, float* src, int8_t* dst ) {
+static void clamp( test_case_Context *ctxt, const float* src, int8_t* dst ) {
 float l;
 float h;
 l = -128.0;
@@ -214,7 +210,7 @@ gemmini_extended_config_st({dst_stride}, {act}, {scale}[0]);
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_17( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale ) {
+void conv_17( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale ) {
 gemmini_extended_config_st((128), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -270,11 +266,11 @@ for (int bo = 0; bo < 1; bo++) {
             }
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (0) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (0) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (0) * (128) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (0) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (0) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (0) * (128) + (64) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (4) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((krow) * (3 * 8 * 8 * 16 * 16) + (0) * (8 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -436,11 +432,11 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (1) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (1) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (1) * (128) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 16 * 16) + (1) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (1) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (1) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (1) * (128) + (64) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 16 * 16) + (1) * (8 * 16 * 16) + (4) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((krow) * (3 * 8 * 8 * 16 * 16) + (1) * (8 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -602,11 +598,11 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (2) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (2) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (2) * (128) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 16 * 16) + (2) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (2) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (2) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (2) * (128) + (64) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 16 * 16) + (2) * (8 * 16 * 16) + (4) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((krow) * (3 * 8 * 8 * 16 * 16) + (2) * (8 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (16 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (16));
@@ -800,11 +796,11 @@ for (int bo = 0; bo < 1; bo++) {
             }
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (16) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (16) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (16) * (128) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 12 * 16) + (0) * (8 * 12 * 16) + (0) * (12 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (12) );
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (16) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (16) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (16) * (128) + (64) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 12 * 16) + (0) * (8 * 12 * 16) + (4) * (12 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (12) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((krow) * (3 * 8 * 8 * 16 * 16) + (0) * (8 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (13 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (12));
@@ -966,11 +962,11 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (17) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (17) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (17) * (128) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 12 * 16) + (1) * (8 * 12 * 16) + (0) * (12 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (12) );
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (17) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (17) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (17) * (128) + (64) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 12 * 16) + (1) * (8 * 12 * 16) + (4) * (12 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (12) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((krow) * (3 * 8 * 8 * 16 * 16) + (1) * (8 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (13 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (12));
@@ -1132,11 +1128,11 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
             }
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (18) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (18) * (128) + (0) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (18) * (128) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 12 * 16) + (2) * (8 * 12 * 16) + (0) * (12 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (12) );
           }
           if (orow_i == 0 || krow == 2) {
-            gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (18) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
+            gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (18) * (128) + (64) * (1)], { 128, 1 } }).strides[0]*1, 1.0f, 0, (12), 2);
 gemmini_extended_mvin3( &inp[(bi + 4 * bo) * (30 * 30 * 128) + (krow + orow_i + 14 * orow_o) * (30 * 128) + (18) * (128) + (64) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 8 * 12 * 16) + (2) * (8 * 12 * 16) + (4) * (12 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (12) );
           }
           gemmini_extended_preload((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)w_s)) + ((krow) * (3 * 8 * 8 * 16 * 16) + (2) * (8 * 8 * 16 * 16) + (0) * (8 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), (uint32_t)(&*(int32_t*)((uint64_t)( ((uint32_t)((uint64_t)res)) + ((0) * (13 * 16) + (0) * (16) + (0) * (1))/16))) | 0x40000000, (16), (16), (16), (12));
@@ -1293,7 +1289,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_17_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale ) {
+void conv_17_cpu( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale ) {
 EXO_ASSUME(28 == 30 - 3 + 1);
 for (int b = 0; b < 4; b++) {
   for (int orow = 0; orow < 28; orow++) {
@@ -1340,7 +1336,7 @@ for (int b = 0; b < 4; b++) {
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_3( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale ) {
+void conv_3( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale ) {
 gemmini_extended_config_st((64), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -1378,7 +1374,7 @@ for (int b = 0; b < 4; b++) {
                 }
               }
               if (orow_ii + 7 * orow_io == 0 || krow == 2) {
-                gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (krow + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (kcol + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
+                gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (krow + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (kcol + 16 * ocol_o) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (16), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (krow + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (kcol + 16 * ocol_o) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_ii + 7 * orow_io) * (3 * 4 * 16 * 16) + (kcol) * (4 * 16 * 16) + (0) * (16 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (16) );
               }
               for (int kch_o = 0; kch_o < 4; kch_o++) {
@@ -1420,7 +1416,7 @@ gemmini_extended_compute_preloaded((uint32_t)(&*(int8_t*)((uint64_t)( ((uint32_t
               }
             }
             if (orow_ii + 7 * orow_io == 0 || krow == 2) {
-              gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (58 * 58 * 64) + (krow + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48 + kcol) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
+              gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (58 * 58 * 64) + (krow + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48 + kcol) * (64) + (0) * (1)], { 64, 1 } }).strides[0]*1, 1.0f, 0, (8), 2);
 gemmini_extended_mvin3( &inp[(b) * (58 * 58 * 64) + (krow + orow_ii + 7 * orow_io + 28 * orow_o) * (58 * 64) + (48 + kcol) * (64) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s_1)) + ((krow + orow_ii + 7 * orow_io) * (3 * 4 * 8 * 16) + (kcol) * (4 * 8 * 16) + (0) * (8 * 16) + (0) * (16) + (0) * (1))/16))), 16*(4), (8) );
             }
             for (int kch_o = 0; kch_o < 4; kch_o++) {
@@ -1459,7 +1455,7 @@ gemm_free((uint64_t)(i_s));
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_30( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale ) {
+void conv_30( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale ) {
 gemmini_extended_config_st((256), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -1491,7 +1487,7 @@ for (int och_out = 0; och_out < 4; och_out++) {
               }
             }
             if (orow_i == 0 || krow == 2) {
-              gemmini_extended4_config_ld(((struct exo_win_2i8){ (int8_t*)&inp[(b) * (16 * 16 * 256) + (krow + orow_i + 7 * orow_o) * (16 * 256) + (kcol) * (256) + (0) * (1)], { 256, 1 } }).strides[0]*1, 1.0f, 0, (14), 2);
+              gemmini_extended4_config_ld(((struct exo_win_2i8c){ &inp[(b) * (16 * 16 * 256) + (krow + orow_i + 7 * orow_o) * (16 * 256) + (kcol) * (256) + (0) * (1)], { 256, 1 } }).strides[0]*1, 1.0f, 0, (14), 2);
 gemmini_extended_mvin3( &inp[(b) * (16 * 16 * 256) + (krow + orow_i + 7 * orow_o) * (16 * 256) + (kcol) * (256) + (0) * (1)], ((uint64_t) &*(int8_t*)((uint64_t)( ((uint32_t)((uint64_t)i_s)) + ((krow + orow_i) * (3 * 16 * 14 * 16) + (kcol) * (16 * 14 * 16) + (0) * (14 * 16) + (0) * (16) + (0) * (1))/16))), 16*(16), (14) );
             }
             for (int kch_o = 0; kch_o < 16; kch_o++) {
@@ -1527,7 +1523,7 @@ gemm_free((uint64_t)(i_s));
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_30_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale ) {
+void conv_30_cpu( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale ) {
 EXO_ASSUME(14 == 16 - 3 + 1);
 for (int b = 0; b < 4; b++) {
   for (int orow = 0; orow < 14; orow++) {
@@ -1574,7 +1570,7 @@ for (int b = 0; b < 4; b++) {
 //     act : bool,
 //     scale : f32  @DRAM
 // )
-void conv_3_cpu( test_case_Context *ctxt, int8_t* output, int32_t* bias, int8_t* inp, int8_t* weights, bool act, float* scale ) {
+void conv_3_cpu( test_case_Context *ctxt, int8_t* output, const int32_t* bias, const int8_t* inp, const int8_t* weights, bool act, const float* scale ) {
 EXO_ASSUME(56 == 58 - 3 + 1);
 for (int b = 0; b < 4; b++) {
   for (int orow = 0; orow < 56; orow++) {
diff --git a/tests/golden/test_apps/test_gemmini_matmul.txt b/tests/golden/test_apps/test_gemmini_matmul.txt
index 3ca1c6d..ca4a631 100644
--- a/tests/golden/test_apps/test_gemmini_matmul.txt
+++ b/tests/golden/test_apps/test_gemmini_matmul.txt
@@ -54,17 +54,21 @@ typedef struct test_case_Context {
     } ConfigStore;
 
 } test_case_Context;
-struct exo_win_2i32{
-    int32_t *data;
-    int_fast32_t strides[2];
+struct exo_win_2i32c{
+    const int32_t * const data;
+    const int_fast32_t strides[2];
 };
 struct exo_win_2i8{
-    int8_t *data;
-    int_fast32_t strides[2];
+    int8_t * const data;
+    const int_fast32_t strides[2];
 };
-struct exo_win_3i8{
-    int8_t *data;
-    int_fast32_t strides[3];
+struct exo_win_2i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_3i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[3];
 };
 // cpu_matmul_14(
 //     scale : f32  @DRAM,
@@ -73,7 +77,7 @@ struct exo_win_3i8{
 //     B : i8[128,512]  @DRAM,
 //     C : i8[3136,512]  @DRAM
 // )
-void cpu_matmul_14( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void cpu_matmul_14( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // cpu_matmul_16(
 //     scale : f32  @DRAM,
@@ -82,7 +86,7 @@ void cpu_matmul_14( test_case_Context *ctxt, float* scale, bool act, int8_t* A,
 //     B : i8[512,128]  @DRAM,
 //     C : i8[3136,128]  @DRAM
 // )
-void cpu_matmul_16( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void cpu_matmul_16( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // cpu_matmul_27(
 //     scale : f32  @DRAM,
@@ -91,7 +95,7 @@ void cpu_matmul_16( test_case_Context *ctxt, float* scale, bool act, int8_t* A,
 //     B : i8[256,1024]  @DRAM,
 //     C : i8[784,1024]  @DRAM
 // )
-void cpu_matmul_27( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void cpu_matmul_27( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // cpu_matmul_4(
 //     scale : f32  @DRAM,
@@ -100,7 +104,7 @@ void cpu_matmul_27( test_case_Context *ctxt, float* scale, bool act, int8_t* A,
 //     B : i8[64,256]  @DRAM,
 //     C : i8[12544,256]  @DRAM
 // )
-void cpu_matmul_4( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void cpu_matmul_4( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // cpu_matmul_512x512x512(
 //     scale : f32  @DRAM,
@@ -109,7 +113,7 @@ void cpu_matmul_4( test_case_Context *ctxt, float* scale, bool act, int8_t* A, i
 //     B : i8[512,512]  @DRAM,
 //     C : i8[512,512]  @DRAM
 // )
-void cpu_matmul_512x512x512( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void cpu_matmul_512x512x512( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // cpu_matmul_6(
 //     scale : f32  @DRAM,
@@ -118,7 +122,7 @@ void cpu_matmul_512x512x512( test_case_Context *ctxt, float* scale, bool act, in
 //     B : i8[256,64]  @DRAM,
 //     C : i8[12544,64]  @DRAM
 // )
-void cpu_matmul_6( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void cpu_matmul_6( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // matmul_14(
 //     scale : f32  @DRAM,
@@ -127,7 +131,7 @@ void cpu_matmul_6( test_case_Context *ctxt, float* scale, bool act, int8_t* A, i
 //     B : i8[128,512]  @DRAM,
 //     C : i8[3136,512]  @DRAM
 // )
-void matmul_14( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void matmul_14( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // matmul_16(
 //     scale : f32  @DRAM,
@@ -136,7 +140,7 @@ void matmul_14( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8
 //     B : i8[512,128]  @DRAM,
 //     C : i8[3136,128]  @DRAM
 // )
-void matmul_16( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void matmul_16( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // matmul_27(
 //     scale : f32  @DRAM,
@@ -145,7 +149,7 @@ void matmul_16( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8
 //     B : i8[256,1024]  @DRAM,
 //     C : i8[784,1024]  @DRAM
 // )
-void matmul_27( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void matmul_27( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // matmul_4(
 //     scale : f32  @DRAM,
@@ -154,7 +158,7 @@ void matmul_27( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8
 //     B : i8[64,256]  @DRAM,
 //     C : i8[12544,256]  @DRAM
 // )
-void matmul_4( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void matmul_4( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // matmul_512x512x512(
 //     scale : f32  @DRAM,
@@ -163,7 +167,7 @@ void matmul_4( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_
 //     B : i8[512,512]  @DRAM,
 //     C : i8[512,512]  @DRAM
 // )
-void matmul_512x512x512( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void matmul_512x512x512( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 // matmul_6(
 //     scale : f32  @DRAM,
@@ -172,7 +176,7 @@ void matmul_512x512x512( test_case_Context *ctxt, float* scale, bool act, int8_t
 //     B : i8[256,64]  @DRAM,
 //     C : i8[12544,64]  @DRAM
 // )
-void matmul_6( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C );
+void matmul_6( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C );
 
 
 
@@ -184,14 +188,6 @@ void matmul_6( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_
 #include "test_case.h"
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -214,7 +210,7 @@ double _select_(double x, double v, double y, double z) {
 //     src : f32  @DRAM,
 //     dst : i8  @DRAM
 // )
-static void clamp( test_case_Context *ctxt, float* src, int8_t* dst );
+static void clamp( test_case_Context *ctxt, const float* src, int8_t* dst );
 
 
 /* relying on the following instruction..."
@@ -225,7 +221,7 @@ acc_scale(src,dst,scale)
 //     src : f32  @DRAM,
 //     dst : i8  @DRAM
 // )
-static void clamp( test_case_Context *ctxt, float* src, int8_t* dst ) {
+static void clamp( test_case_Context *ctxt, const float* src, int8_t* dst ) {
 float l;
 float h;
 l = -128.0;
@@ -271,7 +267,7 @@ gemmini_extended3_config_ld(0, 1.0f, 0, 0);
 //     B : i8[128,512]  @DRAM,
 //     C : i8[3136,512]  @DRAM
 // )
-void cpu_matmul_14( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void cpu_matmul_14( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 for (int i = 0; i < 3136; i++) {
   for (int j = 0; j < 512; j++) {
     int32_t res;
@@ -308,7 +304,7 @@ for (int i = 0; i < 3136; i++) {
 //     B : i8[512,128]  @DRAM,
 //     C : i8[3136,128]  @DRAM
 // )
-void cpu_matmul_16( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void cpu_matmul_16( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 for (int i = 0; i < 3136; i++) {
   for (int j = 0; j < 128; j++) {
     int32_t res;
@@ -345,7 +341,7 @@ for (int i = 0; i < 3136; i++) {
 //     B : i8[256,1024]  @DRAM,
 //     C : i8[784,1024]  @DRAM
 // )
-void cpu_matmul_27( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void cpu_matmul_27( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 for (int i = 0; i < 784; i++) {
   for (int j = 0; j < 1024; j++) {
     int32_t res;
@@ -382,7 +378,7 @@ for (int i = 0; i < 784; i++) {
 //     B : i8[64,256]  @DRAM,
 //     C : i8[12544,256]  @DRAM
 // )
-void cpu_matmul_4( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void cpu_matmul_4( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 for (int i = 0; i < 12544; i++) {
   for (int j = 0; j < 256; j++) {
     int32_t res;
@@ -419,7 +415,7 @@ for (int i = 0; i < 12544; i++) {
 //     B : i8[512,512]  @DRAM,
 //     C : i8[512,512]  @DRAM
 // )
-void cpu_matmul_512x512x512( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void cpu_matmul_512x512x512( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 for (int i = 0; i < 512; i++) {
   for (int j = 0; j < 512; j++) {
     int32_t res;
@@ -456,7 +452,7 @@ for (int i = 0; i < 512; i++) {
 //     B : i8[256,64]  @DRAM,
 //     C : i8[12544,64]  @DRAM
 // )
-void cpu_matmul_6( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void cpu_matmul_6( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 for (int i = 0; i < 12544; i++) {
   for (int j = 0; j < 64; j++) {
     int32_t res;
@@ -519,7 +515,7 @@ gemmini_extended_mvin( 0, ((uint64_t) &{dst_data}),{m}, {n} );
 //     B : i8[128,512]  @DRAM,
 //     C : i8[3136,512]  @DRAM
 // )
-void matmul_14( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void matmul_14( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 gemmini_extended_config_st((512), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -616,7 +612,7 @@ gemm_acc_free((uint32_t)(res));
 //     B : i8[512,128]  @DRAM,
 //     C : i8[3136,128]  @DRAM
 // )
-void matmul_16( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void matmul_16( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 gemmini_extended_config_st((128), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -713,7 +709,7 @@ gemm_acc_free((uint32_t)(res));
 //     B : i8[256,1024]  @DRAM,
 //     C : i8[784,1024]  @DRAM
 // )
-void matmul_27( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void matmul_27( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 gemmini_extended_config_st((1024), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -806,7 +802,7 @@ gemm_acc_free((uint32_t)(res));
 //     B : i8[64,256]  @DRAM,
 //     C : i8[12544,256]  @DRAM
 // )
-void matmul_4( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void matmul_4( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 gemmini_extended_config_st((256), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -901,7 +897,7 @@ gemm_free((uint64_t)(a));
 //     B : i8[512,512]  @DRAM,
 //     C : i8[512,512]  @DRAM
 // )
-void matmul_512x512x512( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void matmul_512x512x512( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 gemmini_extended_config_st((512), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
@@ -1004,7 +1000,7 @@ gemm_free((uint64_t)(a));
 //     B : i8[256,64]  @DRAM,
 //     C : i8[12544,64]  @DRAM
 // )
-void matmul_6( test_case_Context *ctxt, float* scale, bool act, int8_t* A, int8_t* B, int8_t* C ) {
+void matmul_6( test_case_Context *ctxt, const float* scale, bool act, const int8_t* A, const int8_t* B, int8_t* C ) {
 gemmini_extended_config_st((64), (act), (scale)[0]);
 
 gemmini_extended_config_ex(WS, 0, 0, 0, 1, 0, 0);
diff --git a/tests/golden/test_apps/test_neon_sgemm.txt b/tests/golden/test_apps/test_neon_sgemm.txt
index 0003297..08fb54f 100644
--- a/tests/golden/test_apps/test_neon_sgemm.txt
+++ b/tests/golden/test_apps/test_neon_sgemm.txt
@@ -29,16 +29,22 @@ extern "C" {
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct test_case_Context { 
 
-} test_case_Context;
 struct exo_win_1f32{
-    float *data;
-    int_fast32_t strides[1];
+    float * const data;
+    const int_fast32_t strides[1];
+};
+struct exo_win_1f32c{
+    const float * const data;
+    const int_fast32_t strides[1];
 };
 struct exo_win_2f32{
-    float *data;
-    int_fast32_t strides[2];
+    float * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_2f32c{
+    const float * const data;
+    const int_fast32_t strides[2];
 };
 // sgemm_exo(
 //     M : size,
@@ -48,7 +54,7 @@ struct exo_win_2f32{
 //     B : f32[K,N]  @DRAM,
 //     C : f32[M,N]  @DRAM
 // )
-void sgemm_exo( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, float* A, float* B, float* C );
+void sgemm_exo( void *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, const float* A, const float* B, float* C );
 
 
 
@@ -60,14 +66,6 @@ void sgemm_exo( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fas
 #include "test_case.h"
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -80,7 +78,7 @@ static int8_t _clamp_32to8(int32_t x) {
 //     B : [f32][K,16]  @DRAM,
 //     C : [f32][4,16]  @DRAM
 // )
-static void neon_microkernel( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void neon_microkernel( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 
 /* relying on the following instruction..."
@@ -93,7 +91,7 @@ neon_broadcast_4xf32(dst,src)
 //     B : [f32][K,16]  @DRAM,
 //     C : [f32][4,16]  @DRAM
 // )
-static void neon_microkernel( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void neon_microkernel( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
 EXO_ASSUME(B.strides[1] == 1);
@@ -149,7 +147,7 @@ vst1q_f32(&{dst_data}, {src_data});
 //     B : f32[K,N]  @DRAM,
 //     C : f32[M,N]  @DRAM
 // )
-void sgemm_exo( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, float* A, float* B, float* C ) {
+void sgemm_exo( void *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, const float* A, const float* B, float* C ) {
 EXO_ASSUME(M >= 1);
 EXO_ASSUME(N >= 1);
 EXO_ASSUME(K >= 1);
@@ -173,7 +171,7 @@ for (int ko = 0; ko < ((K) / (64)); ko++) {
       }
       for (int im = 0; im < 16; im++) {
         for (int jm = 0; jm < 4; jm++) {
-          neon_microkernel(ctxt,64,(struct exo_win_2f32){ (float*)&Atile[(4 * im) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&Btile[(0) * (64) + (16 * jm) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(4 * im + 64 * io) * (N) + (16 * jm + 64 * jo) * (1)], { N, 1 } });
+          neon_microkernel(ctxt,64,(struct exo_win_2f32c){ &Atile[(4 * im) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32c){ &Btile[(0) * (64) + (16 * jm) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(4 * im + 64 * io) * (N) + (16 * jm + 64 * jo) * (1)], { N, 1 } });
         }
       }
     }
@@ -185,13 +183,13 @@ for (int ko = 0; ko < ((K) / (64)); ko++) {
   for (int io = 0; io < ((((M) / (4))) / (16)); io++) {
     for (int jm = 0; jm < ((N) / (16)) % 4; jm++) {
       for (int im = 0; im < 16; im++) {
-        neon_microkernel(ctxt,64,(struct exo_win_2f32){ (float*)&A[(4 * im + 64 * io) * (K) + (64 * ko) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B[(64 * ko) * (N) + (16 * (jm + ((((N) / (16))) / (4)) * 4)) * (1)], { N, 1 } },(struct exo_win_2f32){ (float*)&C[(4 * im + 64 * io) * (N) + (16 * (jm + ((((N) / (16))) / (4)) * 4)) * (1)], { N, 1 } });
+        neon_microkernel(ctxt,64,(struct exo_win_2f32c){ &A[(4 * im + 64 * io) * (K) + (64 * ko) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B[(64 * ko) * (N) + (16 * (jm + ((((N) / (16))) / (4)) * 4)) * (1)], { N, 1 } },(struct exo_win_2f32){ &C[(4 * im + 64 * io) * (N) + (16 * (jm + ((((N) / (16))) / (4)) * 4)) * (1)], { N, 1 } });
       }
     }
   }
   for (int jo = 0; jo < ((N) / (16)); jo++) {
     for (int im = 0; im < ((M) / (4)) % 16; im++) {
-      neon_microkernel(ctxt,64,(struct exo_win_2f32){ (float*)&A[(4 * (im + ((((M) / (4))) / (16)) * 16)) * (K) + (64 * ko) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B[(64 * ko) * (N) + (16 * jo) * (1)], { N, 1 } },(struct exo_win_2f32){ (float*)&C[(4 * (im + ((((M) / (4))) / (16)) * 16)) * (N) + (16 * jo) * (1)], { N, 1 } });
+      neon_microkernel(ctxt,64,(struct exo_win_2f32c){ &A[(4 * (im + ((((M) / (4))) / (16)) * 16)) * (K) + (64 * ko) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B[(64 * ko) * (N) + (16 * jo) * (1)], { N, 1 } },(struct exo_win_2f32){ &C[(4 * (im + ((((M) / (4))) / (16)) * 16)) * (N) + (16 * jo) * (1)], { N, 1 } });
     }
   }
 }
diff --git a/tests/golden/test_apps/test_x86_conv.txt b/tests/golden/test_apps/test_x86_conv.txt
index 8cb0e22..f3ba153 100644
--- a/tests/golden/test_apps/test_x86_conv.txt
+++ b/tests/golden/test_apps/test_x86_conv.txt
@@ -29,12 +29,14 @@ extern "C" {
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct test_case_Context { 
 
-} test_case_Context;
 struct exo_win_1f32{
-    float *data;
-    int_fast32_t strides[1];
+    float * const data;
+    const int_fast32_t strides[1];
+};
+struct exo_win_1f32c{
+    const float * const data;
+    const int_fast32_t strides[1];
 };
 // conv_specialized(
 //     inp : f32[5,82,102,128]  @DRAM,
@@ -42,7 +44,7 @@ struct exo_win_1f32{
 //     weights : f32[128,3,3,128]  @DRAM,
 //     bias : f32[128]  @DRAM
 // )
-void conv_specialized( test_case_Context *ctxt, float* inp, float* output, float* weights, float* bias );
+void conv_specialized( void *ctxt, const float* inp, float* output, const float* weights, const float* bias );
 
 
 
@@ -54,14 +56,6 @@ void conv_specialized( test_case_Context *ctxt, float* inp, float* output, float
 #include "test_case.h"
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <immintrin.h>
 #include <stdio.h>
@@ -79,7 +73,7 @@ double _relu_(double x) {
 //     weights : f32[128,3,3,128]  @DRAM,
 //     bias : f32[128]  @DRAM
 // )
-void conv_specialized( test_case_Context *ctxt, float* inp, float* output, float* weights, float* bias ) {
+void conv_specialized( void *ctxt, const float* inp, float* output, const float* weights, const float* bias ) {
 for (int oc_o = 0; oc_o < 2; oc_o++) {
   for (int n = 0; n < 5; n++) {
     for (int oy = 0; oy < 80; oy++) {
diff --git a/tests/golden/test_apps/test_x86_sgemm.txt b/tests/golden/test_apps/test_x86_sgemm.txt
index 8414b87..cc9eeb9 100644
--- a/tests/golden/test_apps/test_x86_sgemm.txt
+++ b/tests/golden/test_apps/test_x86_sgemm.txt
@@ -29,16 +29,22 @@ extern "C" {
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct test_case_Context { 
 
-} test_case_Context;
 struct exo_win_1f32{
-    float *data;
-    int_fast32_t strides[1];
+    float * const data;
+    const int_fast32_t strides[1];
+};
+struct exo_win_1f32c{
+    const float * const data;
+    const int_fast32_t strides[1];
 };
 struct exo_win_2f32{
-    float *data;
-    int_fast32_t strides[2];
+    float * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_2f32c{
+    const float * const data;
+    const int_fast32_t strides[2];
 };
 // sgemm_exo(
 //     M : size,
@@ -48,7 +54,7 @@ struct exo_win_2f32{
 //     B : f32[K,N]  @DRAM,
 //     C : f32[M,N]  @DRAM
 // )
-void sgemm_exo( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, float* A, float* B, float* C );
+void sgemm_exo( void *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, const float* A, const float* B, float* C );
 
 
 
@@ -60,14 +66,6 @@ void sgemm_exo( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fas
 #include "test_case.h"
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <immintrin.h>
 #include <stdio.h>
@@ -84,7 +82,7 @@ static int8_t _clamp_32to8(int32_t x) {
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][M,64]  @DRAM
 // )
-static void bottom_panel_kernel_scheduled( test_case_Context *ctxt, int_fast32_t M, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void bottom_panel_kernel_scheduled( void *ctxt, int_fast32_t M, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // right_panel_kernel_scheduled(
 //     N : size,
@@ -93,7 +91,7 @@ static void bottom_panel_kernel_scheduled( test_case_Context *ctxt, int_fast32_t
 //     B : [f32][K,N]  @DRAM,
 //     C : [f32][6,N]  @DRAM
 // )
-static void right_panel_kernel_scheduled( test_case_Context *ctxt, int_fast32_t N, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void right_panel_kernel_scheduled( void *ctxt, int_fast32_t N, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // sgemm_above_kernel(
 //     M : size,
@@ -103,7 +101,7 @@ static void right_panel_kernel_scheduled( test_case_Context *ctxt, int_fast32_t
 //     B : [f32][K,N]  @DRAM,
 //     C : [f32][M,N]  @DRAM
 // )
-static void sgemm_above_kernel( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void sgemm_above_kernel( void *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // sgemm_kernel_avx512_1x4(
 //     K : size,
@@ -111,7 +109,7 @@ static void sgemm_above_kernel( test_case_Context *ctxt, int_fast32_t M, int_fas
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][1,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_1x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void sgemm_kernel_avx512_1x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // sgemm_kernel_avx512_2x4(
 //     K : size,
@@ -119,7 +117,7 @@ static void sgemm_kernel_avx512_1x4( test_case_Context *ctxt, int_fast32_t K, st
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][2,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_2x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void sgemm_kernel_avx512_2x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // sgemm_kernel_avx512_3x4(
 //     K : size,
@@ -127,7 +125,7 @@ static void sgemm_kernel_avx512_2x4( test_case_Context *ctxt, int_fast32_t K, st
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][3,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_3x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void sgemm_kernel_avx512_3x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // sgemm_kernel_avx512_4x4(
 //     K : size,
@@ -135,7 +133,7 @@ static void sgemm_kernel_avx512_3x4( test_case_Context *ctxt, int_fast32_t K, st
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][4,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_4x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void sgemm_kernel_avx512_4x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // sgemm_kernel_avx512_5x4(
 //     K : size,
@@ -143,7 +141,7 @@ static void sgemm_kernel_avx512_4x4( test_case_Context *ctxt, int_fast32_t K, st
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][5,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_5x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void sgemm_kernel_avx512_5x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // sgemm_kernel_avx512_6x4(
 //     K : size,
@@ -151,7 +149,7 @@ static void sgemm_kernel_avx512_5x4( test_case_Context *ctxt, int_fast32_t K, st
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][6,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_6x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C );
+static void sgemm_kernel_avx512_6x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C );
 
 // bottom_panel_kernel_scheduled(
 //     M : size,
@@ -160,7 +158,7 @@ static void sgemm_kernel_avx512_6x4( test_case_Context *ctxt, int_fast32_t K, st
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][M,64]  @DRAM
 // )
-static void bottom_panel_kernel_scheduled( test_case_Context *ctxt, int_fast32_t M, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void bottom_panel_kernel_scheduled( void *ctxt, int_fast32_t M, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(M >= 1);
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
@@ -168,19 +166,19 @@ EXO_ASSUME(B.strides[1] == 1);
 EXO_ASSUME(C.strides[1] == 1);
 EXO_ASSUME(M < 6);
 if (M == 1) {
-  sgemm_kernel_avx512_1x4(ctxt,K,(struct exo_win_2f32){ (float*)&A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+  sgemm_kernel_avx512_1x4(ctxt,K,(struct exo_win_2f32c){ &A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
 } else {
   if (M == 2) {
-    sgemm_kernel_avx512_2x4(ctxt,K,(struct exo_win_2f32){ (float*)&A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+    sgemm_kernel_avx512_2x4(ctxt,K,(struct exo_win_2f32c){ &A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
   } else {
     if (M == 3) {
-      sgemm_kernel_avx512_3x4(ctxt,K,(struct exo_win_2f32){ (float*)&A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+      sgemm_kernel_avx512_3x4(ctxt,K,(struct exo_win_2f32c){ &A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
     } else {
       if (M == 4) {
-        sgemm_kernel_avx512_4x4(ctxt,K,(struct exo_win_2f32){ (float*)&A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+        sgemm_kernel_avx512_4x4(ctxt,K,(struct exo_win_2f32c){ &A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
       } else {
         if (M == 5) {
-          sgemm_kernel_avx512_5x4(ctxt,K,(struct exo_win_2f32){ (float*)&A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+          sgemm_kernel_avx512_5x4(ctxt,K,(struct exo_win_2f32c){ &A.data[(0) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (0) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(0) * (C.strides[0]) + (0) * (C.strides[1])], { C.strides[0], C.strides[1] } });
         } else {
           for (int k = 0; k < K; k++) {
             for (int i = 0; i < M; i++) {
@@ -243,7 +241,7 @@ _mm512_storeu_ps(&{dst_data}, {src_data});
 //     B : [f32][K,N]  @DRAM,
 //     C : [f32][6,N]  @DRAM
 // )
-static void right_panel_kernel_scheduled( test_case_Context *ctxt, int_fast32_t N, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void right_panel_kernel_scheduled( void *ctxt, int_fast32_t N, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(N >= 1);
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
@@ -409,7 +407,7 @@ if (((N) / (16)) == 0) {
 //     B : [f32][K,N]  @DRAM,
 //     C : [f32][M,N]  @DRAM
 // )
-static void sgemm_above_kernel( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void sgemm_above_kernel( void *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(M >= 1);
 EXO_ASSUME(N >= 1);
 EXO_ASSUME(K >= 1);
@@ -418,17 +416,17 @@ EXO_ASSUME(B.strides[1] == 1);
 EXO_ASSUME(C.strides[1] == 1);
 for (int io = 0; io < ((M) / (6)); io++) {
   for (int jo = 0; jo < ((N) / (64)); jo++) {
-    sgemm_kernel_avx512_6x4(ctxt,K,(struct exo_win_2f32){ (float*)&A.data[(6 * io) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (64 * jo) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(6 * io) * (C.strides[0]) + (64 * jo) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+    sgemm_kernel_avx512_6x4(ctxt,K,(struct exo_win_2f32c){ &A.data[(6 * io) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (64 * jo) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(6 * io) * (C.strides[0]) + (64 * jo) * (C.strides[1])], { C.strides[0], C.strides[1] } });
   }
 }
 if (N % 64 > 0) {
   for (int io = 0; io < ((M) / (6)); io++) {
-    right_panel_kernel_scheduled(ctxt,N % 64,K,(struct exo_win_2f32){ (float*)&A.data[(6 * io) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (64 * ((N) / (64))) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(6 * io) * (C.strides[0]) + (64 * ((N) / (64))) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+    right_panel_kernel_scheduled(ctxt,N % 64,K,(struct exo_win_2f32c){ &A.data[(6 * io) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (64 * ((N) / (64))) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(6 * io) * (C.strides[0]) + (64 * ((N) / (64))) * (C.strides[1])], { C.strides[0], C.strides[1] } });
   }
 }
 if (M % 6 > 0) {
   for (int jo = 0; jo < ((N) / (64)); jo++) {
-    bottom_panel_kernel_scheduled(ctxt,M % 6,K,(struct exo_win_2f32){ (float*)&A.data[(6 * ((M) / (6))) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32){ (float*)&B.data[(0) * (B.strides[0]) + (64 * jo) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ (float*)&C.data[(6 * ((M) / (6))) * (C.strides[0]) + (64 * jo) * (C.strides[1])], { C.strides[0], C.strides[1] } });
+    bottom_panel_kernel_scheduled(ctxt,M % 6,K,(struct exo_win_2f32c){ &A.data[(6 * ((M) / (6))) * (A.strides[0]) + (0) * (A.strides[1])], { A.strides[0], A.strides[1] } },(struct exo_win_2f32c){ &B.data[(0) * (B.strides[0]) + (64 * jo) * (B.strides[1])], { B.strides[0], B.strides[1] } },(struct exo_win_2f32){ &C.data[(6 * ((M) / (6))) * (C.strides[0]) + (64 * jo) * (C.strides[1])], { C.strides[0], C.strides[1] } });
   }
   if (N % 64 > 0) {
     for (int k = 0; k < K; k++) {
@@ -450,7 +448,7 @@ if (M % 6 > 0) {
 //     B : f32[K,N]  @DRAM,
 //     C : f32[M,N]  @DRAM
 // )
-void sgemm_exo( test_case_Context *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, float* A, float* B, float* C ) {
+void sgemm_exo( void *ctxt, int_fast32_t M, int_fast32_t N, int_fast32_t K, const float* A, const float* B, float* C ) {
 EXO_ASSUME(M >= 1);
 EXO_ASSUME(N >= 1);
 EXO_ASSUME(K >= 1);
@@ -472,7 +470,7 @@ for (int ko = 0; ko < ((K) / (512)); ko++) {
           B1_cache[(i0) * (64) + (i1) * (1)] = B[(i0 + 512 * ko) * (N) + (i1 + 64 * jo) * (1)];
         }
       }
-      sgemm_above_kernel(ctxt,264,64,512,(struct exo_win_2f32){ (float*)&A1_cache[(0) * (512) + (0) * (1)], { 512, 1 } },(struct exo_win_2f32){ (float*)&B1_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * io) * (N) + (64 * jo) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,264,64,512,(struct exo_win_2f32c){ &A1_cache[(0) * (512) + (0) * (1)], { 512, 1 } },(struct exo_win_2f32c){ &B1_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * io) * (N) + (64 * jo) * (1)], { N, 1 } });
     }
   }
 }
@@ -485,7 +483,7 @@ if (N % 64 > 0) {
       }
     }
     for (int io = 0; io < ((M) / (264)); io++) {
-      sgemm_above_kernel(ctxt,264,N % 64,512,(struct exo_win_2f32){ (float*)&A[(264 * io) * (K) + (512 * ko) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B2_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * io) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,264,N % 64,512,(struct exo_win_2f32c){ &A[(264 * io) * (K) + (512 * ko) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B2_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * io) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
     }
   }
 }
@@ -498,7 +496,7 @@ if (M % 264 > 0) {
           B3_cache[(i0) * (64) + (i1) * (1)] = B[(i0 + 512 * ko) * (N) + (i1 + 64 * jo) * (1)];
         }
       }
-      sgemm_above_kernel(ctxt,M % 264,64,512,(struct exo_win_2f32){ (float*)&A[(264 * ((M) / (264))) * (K) + (512 * ko) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B3_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * ((M) / (264))) * (N) + (64 * jo) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,M % 264,64,512,(struct exo_win_2f32c){ &A[(264 * ((M) / (264))) * (K) + (512 * ko) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B3_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * ((M) / (264))) * (N) + (64 * jo) * (1)], { N, 1 } });
     }
   }
 }
@@ -511,7 +509,7 @@ if (M % 264 > 0) {
           B4_cache[(i0) * (64) + (i1) * (1)] = B[(i0 + 512 * ko) * (N) + (64 * ((N) / (64)) + i1) * (1)];
         }
       }
-      sgemm_above_kernel(ctxt,M % 264,N % 64,512,(struct exo_win_2f32){ (float*)&A[(264 * ((M) / (264))) * (K) + (512 * ko) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B4_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * ((M) / (264))) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,M % 264,N % 64,512,(struct exo_win_2f32c){ &A[(264 * ((M) / (264))) * (K) + (512 * ko) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B4_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * ((M) / (264))) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
     }
   }
 }
@@ -524,7 +522,7 @@ if (K % 512 > 0) {
           B5_cache[(i0) * (64) + (i1) * (1)] = B[(512 * ((K) / (512)) + i0) * (N) + (i1 + 64 * jo) * (1)];
         }
       }
-      sgemm_above_kernel(ctxt,264,64,K % 512,(struct exo_win_2f32){ (float*)&A[(264 * io) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B5_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * io) * (N) + (64 * jo) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,264,64,K % 512,(struct exo_win_2f32c){ &A[(264 * io) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B5_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * io) * (N) + (64 * jo) * (1)], { N, 1 } });
     }
   }
 }
@@ -537,7 +535,7 @@ if (K % 512 > 0) {
           B6_cache[(i0) * (64) + (i1) * (1)] = B[(512 * ((K) / (512)) + i0) * (N) + (64 * ((N) / (64)) + i1) * (1)];
         }
       }
-      sgemm_above_kernel(ctxt,264,N % 64,K % 512,(struct exo_win_2f32){ (float*)&A[(264 * io) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B6_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * io) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,264,N % 64,K % 512,(struct exo_win_2f32c){ &A[(264 * io) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B6_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * io) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
     }
   }
 }
@@ -550,7 +548,7 @@ if (K % 512 > 0) {
           B7_cache[(i0) * (64) + (i1) * (1)] = B[(512 * ((K) / (512)) + i0) * (N) + (i1 + 64 * jo) * (1)];
         }
       }
-      sgemm_above_kernel(ctxt,M % 264,64,K % 512,(struct exo_win_2f32){ (float*)&A[(264 * ((M) / (264))) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B7_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * ((M) / (264))) * (N) + (64 * jo) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,M % 264,64,K % 512,(struct exo_win_2f32c){ &A[(264 * ((M) / (264))) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B7_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * ((M) / (264))) * (N) + (64 * jo) * (1)], { N, 1 } });
     }
   }
 }
@@ -563,7 +561,7 @@ if (K % 512 > 0) {
           B8_cache[(i0) * (64) + (i1) * (1)] = B[(512 * ((K) / (512)) + i0) * (N) + (64 * ((N) / (64)) + i1) * (1)];
         }
       }
-      sgemm_above_kernel(ctxt,M % 264,N % 64,K % 512,(struct exo_win_2f32){ (float*)&A[(264 * ((M) / (264))) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32){ (float*)&B8_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ (float*)&C[(264 * ((M) / (264))) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
+      sgemm_above_kernel(ctxt,M % 264,N % 64,K % 512,(struct exo_win_2f32c){ &A[(264 * ((M) / (264))) * (K) + (512 * ((K) / (512))) * (1)], { K, 1 } },(struct exo_win_2f32c){ &B8_cache[(0) * (64) + (0) * (1)], { 64, 1 } },(struct exo_win_2f32){ &C[(264 * ((M) / (264))) * (N) + (64 * ((N) / (64))) * (1)], { N, 1 } });
     }
   }
 }
@@ -575,7 +573,7 @@ if (K % 512 > 0) {
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][1,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_1x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void sgemm_kernel_avx512_1x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
 EXO_ASSUME(B.strides[1] == 1);
@@ -610,7 +608,7 @@ for (int i = 0; i < 1; i++) {
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][2,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_2x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void sgemm_kernel_avx512_2x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
 EXO_ASSUME(B.strides[1] == 1);
@@ -645,7 +643,7 @@ for (int i = 0; i < 2; i++) {
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][3,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_3x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void sgemm_kernel_avx512_3x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
 EXO_ASSUME(B.strides[1] == 1);
@@ -680,7 +678,7 @@ for (int i = 0; i < 3; i++) {
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][4,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_4x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void sgemm_kernel_avx512_4x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
 EXO_ASSUME(B.strides[1] == 1);
@@ -715,7 +713,7 @@ for (int i = 0; i < 4; i++) {
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][5,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_5x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void sgemm_kernel_avx512_5x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
 EXO_ASSUME(B.strides[1] == 1);
@@ -750,7 +748,7 @@ for (int i = 0; i < 5; i++) {
 //     B : [f32][K,64]  @DRAM,
 //     C : [f32][6,64]  @DRAM
 // )
-static void sgemm_kernel_avx512_6x4( test_case_Context *ctxt, int_fast32_t K, struct exo_win_2f32 A, struct exo_win_2f32 B, struct exo_win_2f32 C ) {
+static void sgemm_kernel_avx512_6x4( void *ctxt, int_fast32_t K, struct exo_win_2f32c A, struct exo_win_2f32c B, struct exo_win_2f32 C ) {
 EXO_ASSUME(K >= 1);
 EXO_ASSUME(A.strides[1] == 1);
 EXO_ASSUME(B.strides[1] == 1);
diff --git a/tests/golden/test_codegen/test_const_buffer_parameters.txt b/tests/golden/test_codegen/test_const_buffer_parameters.txt
new file mode 100644
index 0000000..672a212
--- /dev/null
+++ b/tests/golden/test_codegen/test_const_buffer_parameters.txt
@@ -0,0 +1,110 @@
+
+#pragma once
+#ifndef TEST_H
+#define TEST_H
+
+#ifdef __cplusplus
+extern "C" {
+#endif
+
+
+#include <stdint.h>
+#include <stdbool.h>
+
+// Compiler feature macros adapted from Hedley (public domain)
+// https://github.com/nemequ/hedley
+
+#if defined(__has_builtin)
+#  define EXO_HAS_BUILTIN(builtin) __has_builtin(builtin)
+#else
+#  define EXO_HAS_BUILTIN(builtin) (0)
+#endif
+
+#if EXO_HAS_BUILTIN(__builtin_assume)
+#  define EXO_ASSUME(expr) __builtin_assume(expr)
+#elif EXO_HAS_BUILTIN(__builtin_unreachable)
+#  define EXO_ASSUME(expr) \
+      ((void)((expr) ? 1 : (__builtin_unreachable(), 1)))
+#else
+#  define EXO_ASSUME(expr) ((void)(expr))
+#endif
+
+
+struct exo_win_1f32{
+    float * const data;
+    const int_fast32_t strides[1];
+};
+struct exo_win_1f32c{
+    const float * const data;
+    const int_fast32_t strides[1];
+};
+// memcpy(
+//     N : size,
+//     A : f32[N]  @DRAM,
+//     B : f32[N]  @DRAM
+// )
+void memcpy( void *ctxt, int_fast32_t N, float* A, const float* B );
+
+// memcpy_ab(
+//     N : size,
+//     A : [f32][N]  @DRAM,
+//     B : [f32][N]  @DRAM
+// )
+void memcpy_ab( void *ctxt, int_fast32_t N, struct exo_win_1f32 A, struct exo_win_1f32c B );
+
+// memcpy_b(
+//     N : size,
+//     A : f32[N]  @DRAM,
+//     B : [f32][N]  @DRAM
+// )
+void memcpy_b( void *ctxt, int_fast32_t N, float* A, struct exo_win_1f32c B );
+
+
+
+#ifdef __cplusplus
+}
+#endif
+#endif  // TEST_H
+
+#include "test.h"
+
+
+
+#include <stdio.h>
+#include <stdlib.h>
+
+
+
+// memcpy(
+//     N : size,
+//     A : f32[N]  @DRAM,
+//     B : f32[N]  @DRAM
+// )
+void memcpy( void *ctxt, int_fast32_t N, float* A, const float* B ) {
+for (int i = 0; i < N; i++) {
+  A[(i) * (1)] = B[(i) * (1)];
+}
+}
+
+// memcpy_ab(
+//     N : size,
+//     A : [f32][N]  @DRAM,
+//     B : [f32][N]  @DRAM
+// )
+void memcpy_ab( void *ctxt, int_fast32_t N, struct exo_win_1f32 A, struct exo_win_1f32c B ) {
+for (int i = 0; i < N; i++) {
+  A.data[(i) * (A.strides[0])] = B.data[(i) * (B.strides[0])];
+}
+}
+
+// memcpy_b(
+//     N : size,
+//     A : f32[N]  @DRAM,
+//     B : [f32][N]  @DRAM
+// )
+void memcpy_b( void *ctxt, int_fast32_t N, float* A, struct exo_win_1f32c B ) {
+for (int i = 0; i < N; i++) {
+  A[(i) * (1)] = B.data[(i) * (B.strides[0])];
+}
+}
+
diff --git a/tests/golden/test_precision/test_good_prec2.txt b/tests/golden/test_precision/test_good_prec2.txt
index bafdb65..558d6d2 100644
--- a/tests/golden/test_precision/test_good_prec2.txt
+++ b/tests/golden/test_precision/test_good_prec2.txt
@@ -20,27 +20,17 @@
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct c_code_str_Context { 
 
-} c_code_str_Context;
 
 // hoge(
 //     n : size,
 //     x : f32[n]  @DRAM,
 //     y : f32[n]  @DRAM
 // )
-void hoge( c_code_str_Context *ctxt, int_fast32_t n, float* x, float* y );
+void hoge( void *ctxt, int_fast32_t n, const float* x, const float* y );
 
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -52,7 +42,7 @@ static int8_t _clamp_32to8(int32_t x) {
 //     y : f32[m]  @DRAM,
 //     r : f32  @DRAM
 // )
-static void dot( c_code_str_Context *ctxt, int_fast32_t m, float* x, float* y, float* r );
+static void dot( void *ctxt, int_fast32_t m, const float* x, const float* y, float* r );
 
 // dot(
 //     m : size,
@@ -60,7 +50,7 @@ static void dot( c_code_str_Context *ctxt, int_fast32_t m, float* x, float* y, f
 //     y : f32[m]  @DRAM,
 //     r : f32  @DRAM
 // )
-static void dot( c_code_str_Context *ctxt, int_fast32_t m, float* x, float* y, float* r ) {
+static void dot( void *ctxt, int_fast32_t m, const float* x, const float* y, float* r ) {
 *r = 0.0;
 for (int i = 0; i < m; i++) {
   *r += x[(i) * (1)] * y[(i) * (1)];
@@ -72,7 +62,7 @@ for (int i = 0; i < m; i++) {
 //     x : f32[n]  @DRAM,
 //     y : f32[n]  @DRAM
 // )
-void hoge( c_code_str_Context *ctxt, int_fast32_t n, float* x, float* y ) {
+void hoge( void *ctxt, int_fast32_t n, const float* x, const float* y ) {
 float xy;
 dot(ctxt,n,x,y,&xy);
 }
diff --git a/tests/golden/test_schedules/test_expand_dim3.txt b/tests/golden/test_schedules/test_expand_dim3.txt
index f1522fe..11918c7 100644
--- a/tests/golden/test_schedules/test_expand_dim3.txt
+++ b/tests/golden/test_schedules/test_expand_dim3.txt
@@ -20,27 +20,17 @@
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct c_code_str_Context { 
 
-} c_code_str_Context;
 
 // foo(
 //     n : size,
 //     m : size,
 //     x : i8  @DRAM
 // )
-void foo( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, int8_t* x );
+void foo( void *ctxt, int_fast32_t n, int_fast32_t m, int8_t* x );
 
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -52,7 +42,7 @@ static int8_t _clamp_32to8(int32_t x) {
 //     m : size,
 //     x : i8  @DRAM
 // )
-void foo( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, int8_t* x ) {
+void foo( void *ctxt, int_fast32_t n, int_fast32_t m, int8_t* x ) {
 for (int i = 0; i < n; i++) {
   for (int j = 0; j < m; j++) {
     ; // NO-OP
diff --git a/tests/golden/test_window/test_normalize.txt b/tests/golden/test_window/test_normalize.txt
index 61bb3a1..c3b4bed 100644
--- a/tests/golden/test_window/test_normalize.txt
+++ b/tests/golden/test_window/test_normalize.txt
@@ -20,12 +20,10 @@
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct c_code_str_Context { 
 
-} c_code_str_Context;
-struct exo_win_1f32{
-    float *data;
-    int_fast32_t strides[1];
+struct exo_win_1f32c{
+    const float * const data;
+    const int_fast32_t strides[1];
 };
 // proj(
 //     n : size,
@@ -33,18 +31,10 @@ struct exo_win_1f32{
 //     x : f32[n,m]  @DRAM,
 //     y : f32[m,n]  @DRAM
 // )
-void proj( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, float* x, float* y );
+void proj( void *ctxt, int_fast32_t n, int_fast32_t m, const float* x, const float* y );
 
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -56,7 +46,7 @@ static int8_t _clamp_32to8(int32_t x) {
 //     y : [f32][m]  @DRAM,
 //     r : f32  @DRAM
 // )
-static void dot( c_code_str_Context *ctxt, int_fast32_t m, struct exo_win_1f32 x, struct exo_win_1f32 y, float* r );
+static void dot( void *ctxt, int_fast32_t m, struct exo_win_1f32c x, struct exo_win_1f32c y, float* r );
 
 // dot(
 //     m : size,
@@ -64,7 +54,7 @@ static void dot( c_code_str_Context *ctxt, int_fast32_t m, struct exo_win_1f32 x
 //     y : [f32][m]  @DRAM,
 //     r : f32  @DRAM
 // )
-static void dot( c_code_str_Context *ctxt, int_fast32_t m, struct exo_win_1f32 x, struct exo_win_1f32 y, float* r ) {
+static void dot( void *ctxt, int_fast32_t m, struct exo_win_1f32c x, struct exo_win_1f32c y, float* r ) {
 *r = 0.0;
 for (int i = 0; i < m; i++) {
   *r += x.data[(i) * (x.strides[0])] * y.data[(i) * (y.strides[0])];
@@ -77,12 +67,12 @@ for (int i = 0; i < m; i++) {
 //     x : f32[n,m]  @DRAM,
 //     y : f32[m,n]  @DRAM
 // )
-void proj( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, float* x, float* y ) {
+void proj( void *ctxt, int_fast32_t n, int_fast32_t m, const float* x, const float* y ) {
 EXO_ASSUME(n > 4);
 EXO_ASSUME(m > 4);
 float xy;
 float y2;
-dot(ctxt,m,(struct exo_win_1f32){ (float*)&x[(1) * (m) + (0) * (1)], { 1 } },(struct exo_win_1f32){ (float*)&y[(0) * (n) + (2) * (1)], { n } },&xy);
-dot(ctxt,m,(struct exo_win_1f32){ (float*)&y[(0) * (n) + (3) * (1)], { n } },(struct exo_win_1f32){ (float*)&y[(0) * (n) + (3) * (1)], { n } },&y2);
+dot(ctxt,m,(struct exo_win_1f32c){ &x[(1) * (m) + (0) * (1)], { 1 } },(struct exo_win_1f32c){ &y[(0) * (n) + (2) * (1)], { n } },&xy);
+dot(ctxt,m,(struct exo_win_1f32c){ &y[(0) * (n) + (3) * (1)], { n } },(struct exo_win_1f32c){ &y[(0) * (n) + (3) * (1)], { n } },&y2);
 }
 
diff --git a/tests/golden/test_window/test_stride_assert.txt b/tests/golden/test_window/test_stride_assert.txt
index 253ed7c..91a4ddd 100644
--- a/tests/golden/test_window/test_stride_assert.txt
+++ b/tests/golden/test_window/test_stride_assert.txt
@@ -20,12 +20,14 @@
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct c_code_str_Context { 
 
-} c_code_str_Context;
 struct exo_win_2i8{
-    int8_t *data;
-    int_fast32_t strides[2];
+    int8_t * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_2i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[2];
 };
 // stride_assert(
 //     n : size,
@@ -33,18 +35,10 @@ struct exo_win_2i8{
 //     src : [i8][n,m]  @DRAM,
 //     dst : [i8][n,16]  @DRAM
 // )
-void stride_assert( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8 src, struct exo_win_2i8 dst );
-
+void stride_assert( void *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8c src, struct exo_win_2i8 dst );
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
 
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -57,7 +51,7 @@ static int8_t _clamp_32to8(int32_t x) {
 //     src : [i8][n,m]  @DRAM,
 //     dst : [i8][n,16]  @DRAM
 // )
-void stride_assert( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8 src, struct exo_win_2i8 dst ) {
+void stride_assert( void *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8c src, struct exo_win_2i8 dst ) {
 EXO_ASSUME(n <= 16);
 EXO_ASSUME(m <= 16);
 EXO_ASSUME(src.strides[1] == 1);
diff --git a/tests/golden/test_window/test_window.txt b/tests/golden/test_window/test_window.txt
index 7fd7a40..1cfad3a 100644
--- a/tests/golden/test_window/test_window.txt
+++ b/tests/golden/test_window/test_window.txt
@@ -20,12 +20,14 @@
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct c_code_str_Context { 
 
-} c_code_str_Context;
 struct exo_win_2i8{
-    int8_t *data;
-    int_fast32_t strides[2];
+    int8_t * const data;
+    const int_fast32_t strides[2];
+};
+struct exo_win_2i8c{
+    const int8_t * const data;
+    const int_fast32_t strides[2];
 };
 // window(
 //     n : size,
@@ -33,18 +35,10 @@ struct exo_win_2i8{
 //     src : [i8][n,m]  @DRAM,
 //     dst : [i8][n,16]  @DRAM
 // )
-void window( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8 src, struct exo_win_2i8 dst );
-
+void window( void *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8c src, struct exo_win_2i8 dst );
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
 
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -57,7 +51,7 @@ static int8_t _clamp_32to8(int32_t x) {
 //     src : [i8][n,m]  @DRAM,
 //     dst : [i8][n,16]  @DRAM
 // )
-void window( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8 src, struct exo_win_2i8 dst ) {
+void window( void *ctxt, int_fast32_t n, int_fast32_t m, struct exo_win_2i8c src, struct exo_win_2i8 dst ) {
 EXO_ASSUME(n <= 16);
 EXO_ASSUME(m <= 16);
 for (int i = 0; i < n; i++) {
diff --git a/tests/golden/test_window/test_window_stmt.txt b/tests/golden/test_window/test_window_stmt.txt
index 55cd287..66e630a 100644
--- a/tests/golden/test_window/test_window_stmt.txt
+++ b/tests/golden/test_window/test_window_stmt.txt
@@ -20,30 +20,20 @@
 #  define EXO_ASSUME(expr) ((void)(expr))
 #endif
 
-typedef struct c_code_str_Context { 
 
-} c_code_str_Context;
-struct exo_win_1f32{
-    float *data;
-    int_fast32_t strides[1];
+struct exo_win_1f32c{
+    const float * const data;
+    const int_fast32_t strides[1];
 };
 // window_stmt(
 //     n : size,
 //     m : size,
 //     x : f32[n,m]  @DRAM
 // )
-void window_stmt( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, float* x );
+void window_stmt( void *ctxt, int_fast32_t n, int_fast32_t m, const float* x );
 
 
 
-static int _floor_div(int num, int quot) {
-  int off = (num>=0)? 0 : quot-1;
-  return (num-off)/quot;
-}
-
-static int8_t _clamp_32to8(int32_t x) {
-  return (x < -128)? -128 : ((x > 127)? 127 : x);
-}
 
 #include <stdio.h>
 #include <stdlib.h>
@@ -55,8 +45,8 @@ static int8_t _clamp_32to8(int32_t x) {
 //     m : size,
 //     x : f32[n,m]  @DRAM
 // )
-void window_stmt( c_code_str_Context *ctxt, int_fast32_t n, int_fast32_t m, float* x ) {
-struct exo_win_1f32 y = (struct exo_win_1f32){ (float*)&x[(0) * (m) + (0) * (1)], { m } };
+void window_stmt( void *ctxt, int_fast32_t n, int_fast32_t m, const float* x ) {
+struct exo_win_1f32c y = (struct exo_win_1f32c){ &x[(0) * (m) + (0) * (1)], { m } };
 float *z = malloc(n * sizeof(*z));
 for (int i = 0; i < n; i++) {
   z[(i) * (1)] = y.data[(i) * (y.strides[0])];
diff --git a/tests/test_codegen.py b/tests/test_codegen.py
index 9a22e31..ef13307 100644
--- a/tests/test_codegen.py
+++ b/tests/test_codegen.py
@@ -2,77 +2,88 @@ from __future__ import annotations
 
 from pathlib import Path
 
-import pytest
 import numpy as np
+import pytest
 from PIL import Image
 
-from exo import proc, Procedure, DRAM
+from exo import proc, Procedure, DRAM, compile_procs_to_strings
 from exo.libs.memories import MDRAM, MemGenError
 from exo.stdlib.scheduling import *
 
-
 mock_registers = 0
+
+
 class MOCK(DRAM):
     @classmethod
     def alloc(cls, new_name, prim_type, shape, srcinfo):
         assert len(shape) == 1 and int(shape[0]) == 16
         global mock_registers
         if mock_registers > 0:
-            raise MemGenError('Cannot allocate more than one mock register')
+            raise MemGenError("Cannot allocate more than one mock register")
         mock_registers += 1
 
-        return f'static {prim_type} {new_name}[16];'
+        return f"static {prim_type} {new_name}[16];"
 
     @classmethod
     def free(cls, new_name, prim_type, shape, srcinfo):
         global mock_registers
         mock_registers -= 1
-        return ''
+        return ""
+
 
 # Testing to make sure free is inserted correcctly
 def test_free(compiler):
     @proc
     def foo():
-        x : f32[16] @ MOCK
+        x: f32[16] @ MOCK
         for i in seq(0, 16):
             x[i] = 3.0
-        y : f32[16] @ MOCK
+        y: f32[16] @ MOCK
         for i in seq(0, 16):
             y[i] = 0.0
 
     compiler.compile(foo)
 
+
 def test_free2(compiler):
     @proc
     def foo():
-        x : f32[16] @ MOCK
+        x: f32[16] @ MOCK
         for i in seq(0, 16):
-            y : R[16] @ MOCK
+            y: R[16] @ MOCK
             y[i] = 2.0
 
     compiler.compile(foo)
 
+
 def test_free3(compiler):
     @proc
     def foo():
-        x : f32[16] @ MOCK
+        x: f32[16] @ MOCK
         for i in seq(0, 16):
-            y : R[16] @ MOCK
+            y: R[16] @ MOCK
             x[i] = 2.0
 
     with pytest.raises(MemGenError, match="Cannot allocate"):
         compiler.compile(foo)
 
 
-
 old_split = repeat(divide_loop)
 
+
 # --- Start Blur Test ---
 
+
 def gen_blur() -> Procedure:
     @proc
-    def blur(n: size, m: size, k_size: size,
-             image: R[n, m], kernel: R[k_size, k_size], res: R[n, m]):
+    def blur(
+        n: size,
+        m: size,
+        k_size: size,
+        image: R[n, m],
+        kernel: R[k_size, k_size],
+        res: R[n, m],
+    ):
         for i in seq(0, n):
             for j in seq(0, m):
                 res[i, j] = 0.0
@@ -80,9 +91,13 @@ def gen_blur() -> Procedure:
             for j in seq(0, m):
                 for k in seq(0, k_size):
                     for l in seq(0, k_size):
-                        if i + k >= 1 and i + k - n < 1 and j + l >= 1 and j + l - m < 1:
-                            res[i, j] += kernel[k, l] * image[
-                                i + k - 1, j + l - 1]
+                        if (
+                            i + k >= 1
+                            and i + k - n < 1
+                            and j + l >= 1
+                            and j + l - m < 1
+                        ):
+                            res[i, j] += kernel[k, l] * image[i + k - 1, j + l - 1]
 
     return blur
 
@@ -92,8 +107,7 @@ def _test_blur(compiler, tmp_path, blur):
 
     k_size = 5
 
-    image = np.asarray(Image.open(Path(__file__).parent / 'input.png'),
-                       dtype="float32")
+    image = np.asarray(Image.open(Path(__file__).parent / "input.png"), dtype="float32")
 
     x = np.linspace(-1, 1, k_size + 1)
     kern1d = np.diff(np.random.normal(x))
@@ -105,7 +119,7 @@ def _test_blur(compiler, tmp_path, blur):
     fn(None, *image.shape, k_size, image, kern, res)
 
     out = Image.fromarray(res.astype(np.uint8))
-    out.save(tmp_path / 'out.png')
+    out.save(tmp_path / "out.png")
 
 
 def test_simple_blur(compiler, tmp_path):
@@ -115,9 +129,14 @@ def test_simple_blur(compiler, tmp_path):
 
 def test_simple_blur_split(compiler, tmp_path):
     @proc
-    def simple_blur_split(n: size, m: size, k_size: size,
-                          image: R[n, m], kernel: R[k_size, k_size],
-                          res: R[n, m]):
+    def simple_blur_split(
+        n: size,
+        m: size,
+        k_size: size,
+        image: R[n, m],
+        kernel: R[k_size, k_size],
+        res: R[n, m],
+    ):
         for i in seq(0, n):
             for j1 in seq(0, m / 2):
                 for j2 in seq(0, 2):
@@ -127,9 +146,15 @@ def test_simple_blur_split(compiler, tmp_path):
                 for j2 in seq(0, 2):
                     for k in seq(0, k_size):
                         for l in seq(0, k_size):
-                            if i + k >= 1 and i + k - n < 1 and j1 * 2 + j2 + l >= 1 and j1 * 2 + j2 + l - m < 1:
-                                res[i, j1 * 2 + j2] += kernel[k, l] * image[
-                                    i + k - 1, j1 * 2 + j2 + l - 1]
+                            if (
+                                i + k >= 1
+                                and i + k - n < 1
+                                and j1 * 2 + j2 + l >= 1
+                                and j1 * 2 + j2 + l - m < 1
+                            ):
+                                res[i, j1 * 2 + j2] += (
+                                    kernel[k, l] * image[i + k - 1, j1 * 2 + j2 + l - 1]
+                                )
 
     _test_blur(compiler, tmp_path, simple_blur_split)
 
@@ -137,8 +162,8 @@ def test_simple_blur_split(compiler, tmp_path):
 def test_split_blur(compiler, tmp_path):
     blur = gen_blur()
 
-    blur = old_split(blur, 'j', 4, ['j1', 'j2'])
-    blur = old_split(blur, 'i#1', 4, ['i1', 'i2'])
+    blur = old_split(blur, "j", 4, ["j1", "j2"])
+    blur = old_split(blur, "i#1", 4, ["i1", "i2"])
 
     _test_blur(compiler, tmp_path, blur)
 
@@ -146,8 +171,8 @@ def test_split_blur(compiler, tmp_path):
 def test_reorder_blur(compiler, tmp_path):
     blur = gen_blur()
 
-    blur = reorder_loops(blur, 'k l')
-    blur = reorder_loops(blur, 'i j')
+    blur = reorder_loops(blur, "k l")
+    blur = reorder_loops(blur, "i j")
 
     _test_blur(compiler, tmp_path, blur)
 
@@ -155,8 +180,8 @@ def test_reorder_blur(compiler, tmp_path):
 def test_unroll_blur(compiler, tmp_path):
     blur = gen_blur()
 
-    blur = old_split(blur, 'j', 4, ['j1', 'j2'])
-    blur = repeat(unroll_loop)(blur, 'j2')
+    blur = old_split(blur, "j", 4, ["j1", "j2"])
+    blur = repeat(unroll_loop)(blur, "j2")
 
     _test_blur(compiler, tmp_path, blur)
 
@@ -167,8 +192,7 @@ def test_unroll_blur(compiler, tmp_path):
 # --- conv1d test ---
 def test_conv1d(compiler):
     @proc
-    def conv1d(n: size, m: size, r: size,
-               x: R[n], w: R[m], res: R[r]):
+    def conv1d(n: size, m: size, r: size, x: R[n], w: R[m], res: R[r]):
         for i in seq(0, r):
             res[i] = 0.0
         for i in seq(0, r):
@@ -186,16 +210,19 @@ def test_conv1d(compiler):
     fn = compiler.compile(conv1d)
     fn(None, n_size, m_size, r_size, x, w, res)
 
-    np.testing.assert_almost_equal(res, np.array(
-        [0.12, 0.68, 0.27, -1.26, 2.78, -2.2, 0], dtype=np.float32))
+    np.testing.assert_almost_equal(
+        res, np.array([0.12, 0.68, 0.27, -1.26, 2.78, -2.2, 0], dtype=np.float32)
+    )
 
 
 # ------- Nested alloc test for normal DRAM ------
 
+
 def test_alloc_nest(compiler, tmp_path):
     @proc
-    def alloc_nest(n: size, m: size,
-                   x: R[n, m], y: R[n, m] @ DRAM, res: R[n, m] @ DRAM):
+    def alloc_nest(
+        n: size, m: size, x: R[n, m], y: R[n, m] @ DRAM, res: R[n, m] @ DRAM
+    ):
         for i in seq(0, n):
             rloc: R[m] @ DRAM
             xloc: R[m] @ DRAM
@@ -210,7 +237,7 @@ def test_alloc_nest(compiler, tmp_path):
                 res[i, j] = rloc[j]
 
     # Write effect printing to a file
-    (tmp_path / f'{alloc_nest.name()}_effect.atl').write_text(
+    (tmp_path / f"{alloc_nest.name()}_effect.atl").write_text(
         str(alloc_nest.show_effects())
     )
 
@@ -221,17 +248,19 @@ def test_alloc_nest(compiler, tmp_path):
     fn = compiler.compile(alloc_nest)
     fn(None, *x.shape, x, y, res)
 
-    np.testing.assert_almost_equal(res, np.array(
-        [[3.6, 5.7, 11.9], [4.5, 6.3, 12.0]], dtype=np.float32))
+    np.testing.assert_almost_equal(
+        res, np.array([[3.6, 5.7, 11.9], [4.5, 6.3, 12.0]], dtype=np.float32)
+    )
 
 
 # ------- Nested alloc test for custom malloc DRAM ------
 
+
 def test_alloc_nest_malloc(compiler):
     @proc
-    def alloc_nest_malloc(n: size, m: size,
-                          x: R[n, m] @ MDRAM, y: R[n, m] @ MDRAM,
-                          res: R[n, m] @ MDRAM):
+    def alloc_nest_malloc(
+        n: size, m: size, x: R[n, m] @ MDRAM, y: R[n, m] @ MDRAM, res: R[n, m] @ MDRAM
+    ):
         for i in seq(0, n):
             rloc: R[m] @ MDRAM
             xloc: R[m] @ MDRAM
@@ -250,16 +279,19 @@ def test_alloc_nest_malloc(compiler):
     res = np.zeros_like(x)
 
     root_dir = Path(__file__).parent.parent
-    lib = compiler.compile(alloc_nest_malloc,
-                           include_dir=str(root_dir / "src/exo/libs"),
-                           additional_file=str(root_dir / "src/exo/libs/custom_malloc.c"))
+    lib = compiler.compile(
+        alloc_nest_malloc,
+        include_dir=str(root_dir / "src/exo/libs"),
+        additional_file=str(root_dir / "src/exo/libs/custom_malloc.c"),
+    )
 
     # Initialize custom malloc here
     lib.init_mem()
     lib(None, *x.shape, x, y, res)
 
-    np.testing.assert_almost_equal(res, np.array(
-        [[3.6, 5.7, 11.9], [4.5, 6.3, 12.0]], dtype=np.float32))
+    np.testing.assert_almost_equal(
+        res, np.array([[3.6, 5.7, 11.9], [4.5, 6.3, 12.0]], dtype=np.float32)
+    )
 
 
 def test_unary_neg(compiler):
@@ -351,3 +383,22 @@ def test_select1(compiler):
     fn(None, actual)
 
     np.testing.assert_almost_equal(actual, expected)
+
+
+##
+# Tests for const-correctness
+
+
+def test_const_buffer_parameters(golden, compiler):
+    @proc
+    def memcpy(N: size, A: f32[N], B: f32[N]):
+        for i in seq(0, N):
+            A[i] = B[i]
+
+    memcpy_b = rename(set_window(memcpy, "B", True), "memcpy_b")
+    memcpy_ab = rename(set_window(memcpy_b, "A", True), "memcpy_ab")
+
+    c_file, h_file = compile_procs_to_strings([memcpy, memcpy_b, memcpy_ab], "test.h")
+    code = f"{h_file}\n{c_file}"
+
+    assert code == golden