diff --git a/examples/demo_io/main.bend b/examples/demo_io/main.bend
index b9c22fab..bce1eada 100644
--- a/examples/demo_io/main.bend
+++ b/examples/demo_io/main.bend
@@ -1,63 +1,35 @@
 test-io = 1
 
-type IO_T:
-  Done { magic, expr }
-  Call { magic, func, argm, cont }
+def unwrap(res):
+  match res:
+    case Result/Ok:
+      return res.val
+    case Result/Err:
+      return res.val
 
-def IO_T/MAGIC:
-  return (0xD0CA11, 0xFF1FF1)
+def open():
+  return call("OPEN", ("./LICENSE", "r"))
 
-def IO_T/bind(a, b):
-  match a:
-    case IO_T/Done:
-      return (undefer(b))(a.expr)
-    case IO_T/Call:
-      return IO_T/Call(IO_T/MAGIC, a.func, a.argm, lambda x: IO_T/bind(a.cont(x), b))
+def read(f):
+  return call("READ", (f, 47))
 
-def call_io(func, argm):
-  return IO_T/Call(IO_T/MAGIC, func, argm, lambda x: IO_T/Done(IO_T/MAGIC, x))
+def print(bytes):
+  with IO:
+    * <- call("WRITE", (1, bytes))
+    * <- call("WRITE", (1, "\n"))
 
-def read_input():
-  with IO_T:
-    * <- call_io("WRITE", (1, "What is your name?\n"))
-    return call_io("READ", (0, 10))
+    return wrap(*)
 
-def write_to_file():
-  with IO_T:
-    fp <- call_io("OPEN", ("testing.txt", "w"))
-    input <- read_input()
-    * <- call_io("WRITE", (fp, input))
-    * <- call_io("WRITE", (fp, "\n"))
-
-    return call_io("CLOSE", fp)
-
-def read_from_file():
-  with IO_T:
-    fp <- call_io("OPEN", ("testing.txt", "r"))
-    bytes <- call_io("READ", (fp, 5))
-    * <- call_io("WRITE", (1, bytes))
-    * <- call_io("SEEK", (fp, 2, 0))
-    bytes <- call_io("READ", (fp, 5))
-    * <- call_io("WRITE", (1, bytes))
-
-    return call_io("CLOSE", fp)
-
-def write:
-  return "WRITE"
-
-def one:
-  return 1
-
-def newline:
-  return "\n"
-
-def one_newline_pair:
-  return (one, newline)
+def close(f):
+  return call("CLOSE", f)
 
 def main():
-  with IO_T:
-    * <- write_to_file()
-    * <- read_from_file()
-    * <- call_io(write, one_newline_pair)
-
-    return 42
+  with IO:
+    f <- open()
+    f = unwrap(f)
+    bytes <- read(f)
+    bytes = unwrap(bytes)
+    * <- print(bytes)
+    res <- close(f)
+
+    return wrap(res)
diff --git a/examples/demo_io/main.hvm b/examples/demo_io/main.hvm
index f5f3da3f..c4b72892 100644
--- a/examples/demo_io/main.hvm
+++ b/examples/demo_io/main.hvm
@@ -1,23 +1,26 @@
-@IO_T/Call = (a (b (c (d ((@IO_T/Call/tag (a (b (c (d e))))) e)))))
+@IO/Call = (a (b (c (d ((@IO/Call/tag (a (b (c (d e))))) e)))))
 
-@IO_T/Call/tag = 1
+@IO/Call/tag = 1
 
-@IO_T/Done = (a (b ((@IO_T/Done/tag (a (b c))) c)))
+@IO/Done = (a (b ((@IO/Done/tag (a (b c))) c)))
 
-@IO_T/Done/tag = 0
+@IO/Done/tag = 0
 
-@IO_T/MAGIC = (13683217 16719857)
+@IO/MAGIC = (13683217 16719857)
 
-@IO_T/bind = ((@IO_T/bind__C2 a) a)
+@IO/bind = ((@IO/bind__C2 a) a)
 
-@IO_T/bind__C0 = (* (b (a c)))
+@IO/bind__C0 = (* (b (a c)))
   & @undefer ~ (a (b c))
 
-@IO_T/bind__C1 = (* (* (a (b ((c d) (e g))))))
-  & @IO_T/Call ~ (@IO_T/MAGIC (a (b ((c f) g))))
-  & @IO_T/bind ~ (d (e f))
+@IO/bind__C1 = (* (* (a (b ((c d) (e g))))))
+  & @IO/Call ~ (@IO/MAGIC (a (b ((c f) g))))
+  & @IO/bind ~ (d (e f))
 
-@IO_T/bind__C2 = (?((@IO_T/bind__C0 @IO_T/bind__C1) a) a)
+@IO/bind__C2 = (?((@IO/bind__C0 @IO/bind__C1) a) a)
+
+@IO/wrap = a
+  & @IO/Done ~ (@IO/MAGIC a)
 
 @String/Cons = (a (b ((@String/Cons/tag (a (b c))) c)))
 
@@ -27,215 +30,86 @@
 
 @String/Nil/tag = 0
 
-@call_io = (a (b c))
-  & @IO_T/Call ~ (@IO_T/MAGIC (a (b (@call_io__C0 c))))
-
-@call_io__C0 = a
-  & @IO_T/Done ~ (@IO_T/MAGIC a)
-
-@main = g
-  & @IO_T/bind ~ (@write_to_file ((((* e) f) f) g))
-  & @IO_T/bind ~ (@read_from_file ((((* c) d) d) e))
-  & @IO_T/bind ~ (a ((((* 42) b) b) c))
-  & @call_io ~ (@write (@one_newline_pair a))
-
-@newline = a
-  & @String/Cons ~ (10 (@String/Nil a))
-
-@one = 1
+@call = (a (b c))
+  & @IO/Call ~ (@IO/MAGIC (a (b (@call__C0 c))))
 
-@one_newline_pair = (@one @newline)
+@call__C0 = a
+  & @IO/Done ~ (@IO/MAGIC a)
 
-@read_from_file = a
-  & @IO_T/bind ~ (@read_from_file__C11 (@read_from_file__C10 a))
-
-@read_from_file__C0 = (f (* g))
-  & @call_io ~ (e (f g))
+@close = f
+  & @call ~ (e f)
   & @String/Cons ~ (67 (d e))
   & @String/Cons ~ (76 (c d))
   & @String/Cons ~ (79 (b c))
   & @String/Cons ~ (83 (a b))
   & @String/Cons ~ (69 (@String/Nil a))
 
-@read_from_file__C1 = e
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@read_from_file__C10 = ((@read_from_file__C9 a) a)
-
-@read_from_file__C11 = q
-  & @call_io ~ (d ((o p) q))
+@main = w
+  & @IO/bind ~ (@open ((((s (a u)) (@IO/wrap v)) v) w))
+  & @IO/bind ~ (c ((((n (o (d q))) (r (s t))) t) u))
+  & @unwrap ~ (a {b r})
+  & @read ~ (b c)
+  & @IO/bind ~ (f ((((g (k (* m))) (n (o p))) p) q))
+  & @print ~ (e f)
+  & @unwrap ~ (d e)
+  & @IO/bind ~ (h ((((i i) (k l)) l) m))
+  & @close ~ (g h)
+
+@open = o
+  & @call ~ (d ((m n) o))
   & @String/Cons ~ (79 (c d))
   & @String/Cons ~ (80 (b c))
   & @String/Cons ~ (69 (a b))
   & @String/Cons ~ (78 (@String/Nil a))
-  & @String/Cons ~ (116 (n o))
-  & @String/Cons ~ (101 (m n))
-  & @String/Cons ~ (115 (l m))
-  & @String/Cons ~ (116 (k l))
-  & @String/Cons ~ (105 (j k))
-  & @String/Cons ~ (110 (i j))
-  & @String/Cons ~ (103 (h i))
-  & @String/Cons ~ (46 (g h))
-  & @String/Cons ~ (116 (f g))
-  & @String/Cons ~ (120 (e f))
-  & @String/Cons ~ (116 (@String/Nil e))
-  & @String/Cons ~ (114 (@String/Nil p))
-
-@read_from_file__C2 = (c (a e))
-  & @IO_T/bind ~ (b (((@read_from_file__C0 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C1 ((1 a) b))
-
-@read_from_file__C3 = d
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (65 (a b))
-  & @String/Cons ~ (68 (@String/Nil a))
-
-@read_from_file__C4 = ({a c} (* e))
-  & @IO_T/bind ~ (b (((@read_from_file__C2 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C3 ((a 5) b))
-
-@read_from_file__C5 = d
-  & @String/Cons ~ (83 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (69 (a b))
-  & @String/Cons ~ (75 (@String/Nil a))
-
-@read_from_file__C6 = ({a c} (* e))
-  & @IO_T/bind ~ (b (((@read_from_file__C4 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C5 ((a (2 0)) b))
-
-@read_from_file__C7 = e
+  & @String/Cons ~ (46 (l m))
+  & @String/Cons ~ (47 (k l))
+  & @String/Cons ~ (76 (j k))
+  & @String/Cons ~ (73 (i j))
+  & @String/Cons ~ (67 (h i))
+  & @String/Cons ~ (69 (g h))
+  & @String/Cons ~ (78 (f g))
+  & @String/Cons ~ (83 (e f))
+  & @String/Cons ~ (69 (@String/Nil e))
+  & @String/Cons ~ (114 (@String/Nil n))
+
+@print = (f h)
+  & @IO/bind ~ (g (@print__C3 h))
+  & @call ~ (e ((1 f) g))
   & @String/Cons ~ (87 (d e))
   & @String/Cons ~ (82 (c d))
   & @String/Cons ~ (73 (b c))
   & @String/Cons ~ (84 (a b))
   & @String/Cons ~ (69 (@String/Nil a))
 
-@read_from_file__C8 = (c (a e))
-  & @IO_T/bind ~ (b (((@read_from_file__C6 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C7 ((1 a) b))
-
-@read_from_file__C9 = ({e g} i)
-  & @IO_T/bind ~ (f (((@read_from_file__C8 (g h)) h) i))
-  & @call_io ~ (d ((e 5) f))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (65 (a b))
-  & @String/Cons ~ (68 (@String/Nil a))
-
-@read_input = a
-  & @IO_T/bind ~ (@read_input__C2 (@read_input__C1 a))
+@print__C0 = ((* a) (* a))
 
-@read_input__C0 = (* e)
-  & @call_io ~ (d ((0 10) e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (65 (a b))
-  & @String/Cons ~ (68 (@String/Nil a))
-
-@read_input__C1 = ((@read_input__C0 a) a)
-
-@read_input__C2 = y
-  & @call_io ~ (e ((1 x) y))
+@print__C1 = g
+  & @call ~ (e ((1 f) g))
   & @String/Cons ~ (87 (d e))
   & @String/Cons ~ (82 (c d))
   & @String/Cons ~ (73 (b c))
   & @String/Cons ~ (84 (a b))
   & @String/Cons ~ (69 (@String/Nil a))
-  & @String/Cons ~ (87 (w x))
-  & @String/Cons ~ (104 (v w))
-  & @String/Cons ~ (97 (u v))
-  & @String/Cons ~ (116 (t u))
-  & @String/Cons ~ (32 (s t))
-  & @String/Cons ~ (105 (r s))
-  & @String/Cons ~ (115 (q r))
-  & @String/Cons ~ (32 (p q))
-  & @String/Cons ~ (121 (o p))
-  & @String/Cons ~ (111 (n o))
-  & @String/Cons ~ (117 (m n))
-  & @String/Cons ~ (114 (l m))
-  & @String/Cons ~ (32 (k l))
-  & @String/Cons ~ (110 (j k))
-  & @String/Cons ~ (97 (i j))
-  & @String/Cons ~ (109 (h i))
-  & @String/Cons ~ (101 (g h))
-  & @String/Cons ~ (63 (f g))
   & @String/Cons ~ (10 (@String/Nil f))
 
-@test-io = 1
+@print__C2 = (a (* c))
+  & @IO/bind ~ (@print__C1 (((@print__C0 (a b)) b) c))
 
-@undefer = (((a a) b) b)
-
-@write = e
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@write_to_file = a
-  & @IO_T/bind ~ (@write_to_file__C8 (@write_to_file__C7 a))
-
-@write_to_file__C0 = (f (* g))
-  & @call_io ~ (e (f g))
-  & @String/Cons ~ (67 (d e))
-  & @String/Cons ~ (76 (c d))
-  & @String/Cons ~ (79 (b c))
-  & @String/Cons ~ (83 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
+@print__C3 = ((@print__C2 (@IO/wrap a)) a)
 
-@write_to_file__C1 = a
-  & @String/Cons ~ (10 (@String/Nil a))
-
-@write_to_file__C2 = e
-  & @String/Cons ~ (87 (d e))
+@read = (e f)
+  & @call ~ (d ((e 47) f))
   & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@write_to_file__C3 = ({a c} (* e))
-  & @IO_T/bind ~ (b (((@write_to_file__C0 (c d)) d) e))
-  & @call_io ~ (@write_to_file__C2 ((a @write_to_file__C1) b))
-
-@write_to_file__C4 = e
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
+  & @String/Cons ~ (69 (b c))
+  & @String/Cons ~ (65 (a b))
+  & @String/Cons ~ (68 (@String/Nil a))
 
-@write_to_file__C5 = ({a d} (b f))
-  & @IO_T/bind ~ (c (((@write_to_file__C3 (d e)) e) f))
-  & @call_io ~ (@write_to_file__C4 ((a b) c))
+@test-io = 1
 
-@write_to_file__C6 = (a c)
-  & @IO_T/bind ~ (@read_input (((@write_to_file__C5 (a b)) b) c))
+@undefer = (((a a) b) b)
 
-@write_to_file__C7 = ((@write_to_file__C6 a) a)
+@unwrap = ((@unwrap__C0 a) a)
 
-@write_to_file__C8 = q
-  & @call_io ~ (d ((o p) q))
-  & @String/Cons ~ (79 (c d))
-  & @String/Cons ~ (80 (b c))
-  & @String/Cons ~ (69 (a b))
-  & @String/Cons ~ (78 (@String/Nil a))
-  & @String/Cons ~ (116 (n o))
-  & @String/Cons ~ (101 (m n))
-  & @String/Cons ~ (115 (l m))
-  & @String/Cons ~ (116 (k l))
-  & @String/Cons ~ (105 (j k))
-  & @String/Cons ~ (110 (i j))
-  & @String/Cons ~ (103 (h i))
-  & @String/Cons ~ (46 (g h))
-  & @String/Cons ~ (116 (f g))
-  & @String/Cons ~ (120 (e f))
-  & @String/Cons ~ (116 (@String/Nil e))
-  & @String/Cons ~ (119 (@String/Nil p))
+@unwrap__C0 = (?(((a a) (* (b b))) c) c)
 
 
diff --git a/src/hvm.c b/src/hvm.c
index 8ef9f407..8d116de9 100644
--- a/src/hvm.c
+++ b/src/hvm.c
@@ -1234,7 +1234,7 @@ Port expand(Net* net, Book* book, Port port) {
   Port old = vars_load(net, get_val(ROOT));
   Port got = peek(net, port);
   while (get_tag(got) == REF) {
-    boot_redex(net, new_pair(new_port(REF,get_val(got)), ROOT));
+    boot_redex(net, new_pair(got, ROOT));
     normalize(net, book);
     got = peek(net, vars_load(net, get_val(ROOT)));
   }
diff --git a/src/hvm.cu b/src/hvm.cu
index 1ddb631b..ffa3401b 100644
--- a/src/hvm.cu
+++ b/src/hvm.cu
@@ -1714,7 +1714,7 @@ __global__ void evaluator(GNet* gnet) {
     u32 bag  = tm.mode == SEED ? transpose(GID(), TPB, BPG) : GID();
     u32 rpos = gnet->rbag_pos[bag];
     for (tick = 0; tick < 1 << 9; ++tick) {
-      if (tm.rbag.lo_end > rpos) {
+      if (tm.rbag.lo_end > rpos || rbag_has_highs(&tm.rbag)) {
         if (interact(&net, &tm, pop_redex(&tm), gnet->turn)) {
           while (rbag_has_highs(&tm.rbag)) {
             if (!interact(&net, &tm, pop_redex(&tm), gnet->turn)) break;
@@ -1873,7 +1873,7 @@ Port gnet_expand(GNet* gnet, Port port) {
   Port got = gnet_peek(gnet, port);
   //printf("expand %s\n", show_port(got).x);
   while (get_tag(got) == REF) {
-    gnet_boot_redex(gnet, new_pair(new_port(REF,get_val(got)), ROOT));
+    gnet_boot_redex(gnet, new_pair(got, ROOT));
     gnet_normalize(gnet);
     got = gnet_peek(gnet, gnet_vars_load(gnet, get_val(ROOT)));
   }
diff --git a/src/hvm.cuh b/src/hvm.cuh
index f4d73ea2..1c07a6c3 100644
--- a/src/hvm.cuh
+++ b/src/hvm.cuh
@@ -194,8 +194,8 @@ typedef struct Tup {
 extern Tup gnet_readback_tup(GNet* gnet, Port port, u32 size);
 
 typedef struct Str {
-  u32  text_len;
-  char text_buf[256];
+  u32  len;
+  char *buf;
 } Str;
 
 // Reads a constructor-encoded string (of length at most 255 characters),
diff --git a/src/hvm.h b/src/hvm.h
index b4930e25..b9b59500 100644
--- a/src/hvm.h
+++ b/src/hvm.h
@@ -206,8 +206,8 @@ typedef struct Tup {
 extern Tup readback_tup(Net* net, Book* book, Port port, u32 size);
 
 typedef struct Str {
-  u32  text_len;
-  char text_buf[256];
+  u32  len;
+  char *buf;
 } Str;
 
 // Reads a constructor-encoded string (of length at most 255 characters),
diff --git a/src/run.c b/src/run.c
index b072a1fd..13b3fa94 100644
--- a/src/run.c
+++ b/src/run.c
@@ -1,4 +1,6 @@
 #include <dlfcn.h>
+#include <errno.h>
+#include <stdio.h>
 #include "hvm.c"
 
 // Readback: λ-Encoded Ctr
@@ -35,7 +37,25 @@ typedef struct Bytes {
 #define IO_DONE 0
 #define IO_CALL 1
 
-// List Type
+// Result Tags = Result<T, E>
+#define RESULT_OK  0
+#define RESULT_ERR 1
+
+// IOError = {
+//   Type,           -- a type error
+//   Name,           -- invalid io func name
+//   Inner {val: T}, -- an error while calling an io func
+// }
+#define IO_ERR_TYPE 0
+#define IO_ERR_NAME 1
+#define IO_ERR_INNER 2
+
+typedef struct IOError {
+  u32 tag;
+  Port val;
+} IOError;
+
+// List Tags
 #define LIST_NIL  0
 #define LIST_CONS 1
 
@@ -225,6 +245,134 @@ Port inject_bytes(Net* net, Bytes *bytes) {
   return port;
 }
 
+/// Returns a λ-Encoded Ctr for a RESULT_OK: λt ((t RESULT_OK) val)
+Port inject_ok(Net* net, Port val) {
+  if (!get_resources(net, tm[0], 0, 3, 1)) {
+    fprintf(stderr, "inject_ok: failed to get resources\n");
+    return new_port(ERA, 0);
+  }
+
+  u32 v1 = tm[0]->vloc[0];
+
+  u32 n1 = tm[0]->nloc[0];
+  u32 n2 = tm[0]->nloc[1];
+  u32 n3 = tm[0]->nloc[2];
+
+  vars_create(net, v1, NONE);
+  Port var = new_port(VAR, v1);
+
+  node_create(net, n1, new_pair(val, var));
+  node_create(net, n2, new_pair(new_port(NUM, new_u24(RESULT_OK)), new_port(CON, n1)));
+  node_create(net, n3, new_pair(new_port(CON, n2), var));
+
+  return new_port(CON, n3);
+}
+
+/// Returns a λ-Encoded Ctr for a RESULT_ERR: λt ((t RESULT_ERR) err)
+Port inject_err(Net* net, Port err) {
+  if (!get_resources(net, tm[0], 0, 3, 1)) {
+    fprintf(stderr, "inject_err: failed to get resources\n");
+    return new_port(ERA, 0);
+  }
+
+  u32 v1 = tm[0]->vloc[0];
+
+  u32 n1 = tm[0]->nloc[0];
+  u32 n2 = tm[0]->nloc[1];
+  u32 n3 = tm[0]->nloc[2];
+
+  vars_create(net, v1, NONE);
+  Port var = new_port(VAR, v1);
+
+  node_create(net, n1, new_pair(err, var));
+  node_create(net, n2, new_pair(new_port(NUM, new_u24(RESULT_ERR)), new_port(CON, n1)));
+  node_create(net, n3, new_pair(new_port(CON, n2), var));
+
+  return new_port(CON, n3);
+}
+
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError(..))
+Port inject_io_err(Net* net, IOError err) {
+  if (err.tag <= IO_ERR_NAME) {
+    if (!get_resources(net, tm[0], 0, 2, 1)) {
+      fprintf(stderr, "inject_io_err: failed to get resources\n");
+      return new_port(ERA, 0);
+    }
+
+    u32 v1 = tm[0]->vloc[0];
+
+    u32 n1 = tm[0]->nloc[0];
+    u32 n2 = tm[0]->nloc[1];
+
+    vars_create(net, v1, NONE);
+    Port var = new_port(VAR, v1);
+
+    node_create(net, n1, new_pair(new_port(NUM, new_u24(err.tag)), var));
+    node_create(net, n2, new_pair(new_port(CON, n1), var));
+
+    return inject_err(net, new_port(CON, n2));
+  }
+
+  if (!get_resources(net, tm[0], 0, 3, 1)) {
+    fprintf(stderr, "inject_io_err: failed to get resources\n");
+    return new_port(ERA, 0);
+  }
+
+  u32 v1 = tm[0]->vloc[0];
+
+  u32 n1 = tm[0]->nloc[0];
+  u32 n2 = tm[0]->nloc[1];
+  u32 n3 = tm[0]->nloc[2];
+
+  vars_create(net, v1, NONE);
+  Port var = new_port(VAR, v1);
+
+  node_create(net, n1, new_pair(err.val, var));
+  node_create(net, n2, new_pair(new_port(NUM, new_u24(IO_ERR_INNER)), new_port(CON, n1)));
+  node_create(net, n3, new_pair(new_port(CON, n2), var));
+
+  return inject_err(net, new_port(CON, n3));
+}
+
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError/Type)
+Port inject_io_err_type(Net* net) {
+  IOError io_error = {
+    .tag = IO_ERR_TYPE,
+  };
+
+  return inject_io_err(net, io_error);
+}
+
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError/Name)
+Port inject_io_err_name(Net* net) {
+  IOError io_error = {
+    .tag = IO_ERR_NAME,
+  };
+
+  return inject_io_err(net, io_error);
+}
+
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError/Inner(val))
+Port inject_io_err_inner(Net* net, Port val) {
+  IOError io_error = {
+    .tag = IO_ERR_INNER,
+    .val = val,
+  };
+
+  return inject_io_err(net, io_error);
+}
+
+/// Returns a λ-Encoded Ctr for an Result<T, IOError<String>>
+/// `err` must be `NUL`-terminated.
+Port inject_io_err_str(Net* net, char* err) {
+  Bytes err_bytes;
+  err_bytes.buf = err;
+  err_bytes.len = strlen(err_bytes.buf);
+  Port err_port = inject_bytes(net, &err_bytes);
+
+  return inject_io_err_inner(net, err_port);
+}
+
 // Primitive IO Fns
 // -----------------
 
@@ -255,7 +403,6 @@ FILE* readback_file(Port port) {
 
   FILE* fp = FILE_POINTERS[idx];
   if (fp == NULL) {
-    fprintf(stderr, "invalid file descriptor\n");
     return NULL;
   }
 
@@ -282,19 +429,18 @@ void* readback_dylib(Port port) {
 
 // Reads from a file a specified number of bytes.
 // `argm` is a tuple of (file_descriptor, num_bytes).
+// Returns: Result<Bytes, IOError<i24>>
 Port io_read(Net* net, Book* book, Port argm) {
   Tup tup = readback_tup(net, book, argm, 2);
   if (tup.elem_len != 2) {
-    fprintf(stderr, "io_read: expected 2-tuple\n");
-    return new_port(ERA, 0);
+    return inject_io_err_type(net);
   }
 
   FILE* fp = readback_file(tup.elem_buf[0]);
   u32 num_bytes = get_u24(get_val(tup.elem_buf[1]));
 
   if (fp == NULL) {
-    fprintf(stderr, "io_read: invalid file descriptor\n");
-    return new_port(ERA, 0);
+    return inject_io_err_inner(net, new_port(NUM, new_i24(EBADF)));
   }
 
   /// Read a string.
@@ -303,25 +449,25 @@ Port io_read(Net* net, Book* book, Port argm) {
   bytes.len = fread(bytes.buf, sizeof(char), num_bytes, fp);
 
   if ((bytes.len != num_bytes) && ferror(fp)) {
-    fprintf(stderr, "io_read: failed to read\n");
     free(bytes.buf);
-    return new_port(ERA, 0);
+    return inject_io_err_inner(net, new_port(NUM, new_i24(ferror(fp))));
   }
 
   // Convert it to a port.
   Port ret = inject_bytes(net, &bytes);
   free(bytes.buf);
-  return ret;
+
+  return inject_ok(net, ret);
 }
 
 // Opens a file with the provided mode.
 // `argm` is a tuple (CON node) of the
 // file name and mode as strings.
+// Returns: Result<File, IOError<i24>>
 Port io_open(Net* net, Book* book, Port argm) {
   Tup tup = readback_tup(net, book, argm, 2);
   if (tup.elem_len != 2) {
-    fprintf(stderr, "io_open: expected 2-tuple\n");
-    return new_port(ERA, 0);
+    return inject_io_err_type(net);
   }
 
   Str name = readback_str(net, book, tup.elem_buf[0]);
@@ -334,78 +480,81 @@ Port io_open(Net* net, Book* book, Port argm) {
       free(name.buf);
       free(mode.buf);
 
-      return new_port(NUM, new_u24(fd));
+      if (FILE_POINTERS[fd] == NULL) {
+        return inject_io_err_inner(net, new_port(NUM, new_i24(errno)));
+      }
+
+      return inject_ok(net, new_port(NUM, new_u24(fd)));
     }
   }
 
-  fprintf(stderr, "io_open: too many open files\n");
-
   free(name.buf);
   free(mode.buf);
 
-  return new_port(ERA, 0);
+  // too many open files
+  return inject_io_err_inner(net, new_port(NUM, new_i24(EMFILE)));
 }
 
 // Closes a file, reclaiming the file descriptor.
+// Returns: Result<*, IOError<i24>>
 Port io_close(Net* net, Book* book, Port argm) {
   FILE* fp = readback_file(argm);
   if (fp == NULL) {
-    fprintf(stderr, "io_close: invalid file descriptor\n");
-    return new_port(ERA, 0);
+    return inject_io_err_inner(net, new_port(NUM, new_i24(EBADF)));
   }
 
-  int err = fclose(fp) != 0;
-  if (err != 0) {
-    fprintf(stderr, "io_close: failed to close: %i\n", err);
-    return new_port(ERA, 0);
+  if (fclose(fp) != 0) {
+    return inject_io_err_inner(net, new_port(NUM, new_i24(ferror(fp))));
   }
 
   FILE_POINTERS[get_u24(get_val(argm))] = NULL;
-  return new_port(ERA, 0);
+
+  return inject_ok(net, new_port(ERA, 0));
 }
 
 // Writes a list of bytes to a file.
 // `argm` is a tuple (CON node) of the
 // file descriptor and list of bytes to write.
+// Returns: Result<*, IOError<i24>>
 Port io_write(Net* net, Book* book, Port argm) {
   Tup tup = readback_tup(net, book, argm, 2);
   if (tup.elem_len != 2) {
-    fprintf(stderr, "io_write: expected 2-tuple\n");
-    return new_port(ERA, 0);
+    return inject_io_err_type(net);
   }
 
   FILE* fp = readback_file(tup.elem_buf[0]);
   Bytes bytes = readback_bytes(net, book, tup.elem_buf[1]);
 
   if (fp == NULL) {
-    fprintf(stderr, "io_write: invalid file descriptor\n");
     free(bytes.buf);
-    return new_port(ERA, 0);
+
+    return inject_io_err_inner(net, new_port(NUM, new_i24(EBADF)));
   }
 
   if (fwrite(bytes.buf, sizeof(char), bytes.len, fp) != bytes.len) {
-    fprintf(stderr, "io_write: failed to write\n");
+    free(bytes.buf);
+
+    return inject_io_err_inner(net, new_port(NUM, new_i24(ferror(fp))));
   }
 
   free(bytes.buf);
-  return new_port(ERA, 0);
+
+  return inject_ok(net, new_port(ERA, 0));
 }
 
 // Flushes an output stream.
+// Returns: Result<*, IOError<i24>>
 Port io_flush(Net* net, Book* book, Port argm) {
   FILE* fp = readback_file(argm);
   if (fp == NULL) {
-    fprintf(stderr, "io_flush: invalid file descriptor\n");
-    return new_port(ERA, 0);
+    return inject_io_err_inner(net, new_port(NUM, new_i24(EBADF)));
   }
 
-  int err = fflush(fp) != 0;
-  if (err != 0) {
-    fprintf(stderr, "io_flush: failed to flush: %i\n", err);
-    return new_port(ERA, 0);
+  if (fflush(fp) != 0) {
+    return inject_io_err_inner(net, new_port(NUM, new_i24(ferror(fp))));
   }
 
-  return new_port(ERA, 0);
+  return inject_ok(net, new_port(ERA, 0));
 }
 
 // Seeks to a position in a file.
@@ -416,11 +565,11 @@ Port io_flush(Net* net, Book* book, Port argm) {
 //    - 0 (SEEK_SET): beginning of file
 //    - 1 (SEEK_CUR): current position of the file pointer
 //    - 2 (SEEK_END): end of the file
+// Returns: Result<*, IOError<i24>>
 Port io_seek(Net* net, Book* book, Port argm) {
   Tup tup = readback_tup(net, book, argm, 3);
   if (tup.elem_len != 3) {
-    fprintf(stderr, "io_seek: expected 3-tuple\n");
-    return new_port(ERA, 0);
+    return inject_io_err_type(net);
   }
 
   FILE* fp = readback_file(tup.elem_buf[0]);
@@ -428,8 +577,7 @@ Port io_seek(Net* net, Book* book, Port argm) {
   u32 whence = get_i24(get_val(tup.elem_buf[2]));
 
   if (fp == NULL) {
-    fprintf(stderr, "io_write: invalid file descriptor\n");
-    return new_port(ERA, 0);
+    return inject_io_err_inner(net, new_port(NUM, new_i24(EBADF)));
   }
 
   int cwhence;
@@ -438,19 +586,19 @@ Port io_seek(Net* net, Book* book, Port argm) {
     case 1: cwhence = SEEK_CUR; break;
     case 2: cwhence = SEEK_END; break;
     default:
-      fprintf(stderr, "io_seek: invalid whence\n");
-      return new_port(ERA, 0);
+      return inject_io_err_type(net);
   }
 
   if (fseek(fp, offset, cwhence) != 0) {
-    fprintf(stderr, "io_seek: failed to seek\n");
+    return inject_io_err_inner(net, new_port(NUM, new_i24(ferror(fp))));
   }
 
-  return new_port(ERA, 0);
+  return inject_ok(net, new_port(ERA, 0));
 }
 
 // Returns the current time as a tuple of the high
 // and low 24 bits of a 48-bit nanosecond timestamp.
+// Returns: Result<(u24, u24), IOError<*>>
 Port io_get_time(Net* net, Book* book, Port argm) {
   // Get the current time in nanoseconds
   u64 time_ns = time64();
@@ -461,18 +609,18 @@ Port io_get_time(Net* net, Book* book, Port argm) {
   u32 lps = 0;
   u32 loc = node_alloc_1(net, tm[0], &lps);
   node_create(net, loc, new_pair(new_port(NUM, new_u24(time_hi)), new_port(NUM, new_u24(time_lo))));
-  // Return the encoded time
-  return new_port(CON, loc);
+
+  return inject_ok(net, new_port(CON, loc));
 }
 
 // Sleeps.
 // `argm` is a tuple (CON node) of the high and low
 // 24 bits for a 48-bit duration in nanoseconds.
+// Returns: Result<*, IOError<*>>
 Port io_sleep(Net* net, Book* book, Port argm) {
   Tup tup = readback_tup(net, book, argm, 2);
   if (tup.elem_len != 2) {
-    fprintf(stderr, "io_sleep: expected 2-tuple\n");
-    return new_port(ERA, 0);
+    return inject_io_err_type(net);
   }
 
   // Get the sleep duration node
@@ -487,14 +635,15 @@ Port io_sleep(Net* net, Book* book, Port argm) {
   ts.tv_sec = dur_ns / 1000000000;
   ts.tv_nsec = dur_ns % 1000000000;
   nanosleep(&ts, NULL);
-  // Return an eraser
-  return new_port(ERA, 0);
+
+  return inject_ok(net, new_port(ERA, 0));
 }
 
 // Opens a dylib at the provided path.
 // `argm` is a tuple of `filename` and `lazy`.
 // `filename` is a λ-encoded string.
 // `lazy` is a `bool` indicating if functions should be lazily loaded.
+// Returns: Result<Dylib, IOError<String>>
 Port io_dl_open(Net* net, Book* book, Port argm) {
   Tup tup = readback_tup(net, book, argm, 2);
   Str str = readback_str(net, book, tup.elem_buf[0]);
@@ -505,18 +654,18 @@ Port io_dl_open(Net* net, Book* book, Port argm) {
   for (u32 dl = 0; dl < sizeof(DYLIBS); dl++) {
     if (DYLIBS[dl] == NULL) {
       DYLIBS[dl] = dlopen(str.buf, flags);
-      if (DYLIBS[dl] == NULL) {
-        fprintf(stderr, "failed to open dylib '%s': %s\n", str.buf, dlerror());
 
-        return new_port(ERA, 0);
+      free(str.buf);
+
+      if (DYLIBS[dl] == NULL) {
+        return inject_io_err_str(net, dlerror());
       }
 
-      return new_port(NUM, new_u24(dl));
+      return inject_ok(net, new_port(NUM, new_u24(dl)));
     }
   }
 
-  fprintf(stderr, "io_dl_open: too many open dylibs\n");
-  return new_port(ERA, 0);
+  return inject_io_err_str(net, "too many open dylibs");
 }
 
 // Calls a function from a loaded dylib.
@@ -524,11 +673,15 @@ Port io_dl_open(Net* net, Book* book, Port argm) {
 // `dylib_handle` is the numeric node returned from a `DL_OPEN` call.
 // `symbol` is a λ-encoded string of the symbol name.
 // `args` is the argument to be provided to the dylib symbol.
+//
+// This function returns a Result with an Ok variant containing an
+// arbitrary type.
+//
+// Returns Result<T, IOError<String>>
 Port io_dl_call(Net* net, Book* book, Port argm) {
   Tup tup = readback_tup(net, book, argm, 3);
   if (tup.elem_len != 3) {
-    fprintf(stderr, "io_dl_call: expected 3-tuple\n");
-    return new_port(ERA, 0);
+    return inject_io_err_type(net);
   }
 
   void* dl = readback_dylib(tup.elem_buf[0]);
@@ -538,28 +691,29 @@ Port io_dl_call(Net* net, Book* book, Port argm) {
   Port (*func)(Net*, Book*, Port) = dlsym(dl, symbol.buf);
   char* error = dlerror();
   if (error != NULL) {
-    fprintf(stderr, "io_dl_call: failed to get symbol '%s': %s\n", symbol.buf, error);
+    return inject_io_err_str(net, error);
   }
 
-  return func(net, book, tup.elem_buf[2]);
+  return inject_ok(net, func(net, book, tup.elem_buf[2]));
 }
 
 // Closes a loaded dylib, reclaiming the handle.
+//
+// Returns:  Result<*, IOError<String>>
 Port io_dl_close(Net* net, Book* book, Port argm) {
   void* dl = readback_dylib(argm);
   if (dl == NULL) {
-    fprintf(stderr, "io_dl_close: invalid handle\n");
-    return new_port(ERA, 0);
+    return inject_io_err_type(net);
   }
 
   int err = dlclose(dl) != 0;
   if (err != 0) {
-    fprintf(stderr, "io_dl_close: failed to close: %i\n", err);
-    return new_port(ERA, 0);
+    return inject_io_err_str(net, dlerror());
   }
 
   DYLIBS[get_u24(get_val(argm))] = NULL;
-  return new_port(ERA, 0);
+
+  return inject_ok(net, new_port(ERA, 0));
 }
 
 // Book Loader
@@ -598,7 +752,7 @@ void do_run_io(Net* net, Book* book, Port port) {
     Ctr ctr = readback_ctr(net, book, peek(net, port));
 
     // Checks if IO Magic Number is a CON
-    if (get_tag(ctr.args_buf[0]) != CON) {
+    if (ctr.args_len < 1 || get_tag(ctr.args_buf[0]) != CON) {
       break;
     }
 
@@ -611,6 +765,11 @@ void do_run_io(Net* net, Book* book, Port port) {
 
     switch (ctr.tag) {
       case IO_CALL: {
+        if (ctr.args_len != 4) {
+          fprintf(stderr, "invalid IO_CALL: args_len = %u\n", ctr.args_len);
+          break;
+        }
+
         Str  func = readback_str(net, book, ctr.args_buf[1]);
         FFn* ffn  = NULL;
         // FIXME: optimize this linear search
@@ -620,27 +779,28 @@ void do_run_io(Net* net, Book* book, Port port) {
             break;
           }
         }
-        if (ffn == NULL) {
-          fprintf(stderr, "Unknown IO func '%s'\n", func.buf);
-
-          free(func.buf);
-
-          break;
-        }
 
         free(func.buf);
 
         Port argm = ctr.args_buf[2];
         Port cont = ctr.args_buf[3];
-        Port ret = ffn->func(net, book, argm);
+
+        Port ret;
+        if (ffn == NULL) {
+          ret = inject_io_err_name(net);
+        } else {
+          ret = ffn->func(net, book, argm);
+        };
 
         u32 lps = 0;
         u32 loc = node_alloc_1(net, tm[0], &lps);
         node_create(net, loc, new_pair(ret, ROOT));
         boot_redex(net, new_pair(new_port(CON, loc), cont));
         port = ROOT;
+
         continue;
       }
+
       case IO_DONE: {
         break;
       }
diff --git a/src/run.cu b/src/run.cu
index 3350d9f2..970f8a01 100644
--- a/src/run.cu
+++ b/src/run.cu
@@ -1,4 +1,6 @@
 #include <dlfcn.h>
+#include <errno.h>
+#include <stdio.h>
 #include "hvm.cu"
 
 // Readback: λ-Encoded Ctr
@@ -35,6 +37,24 @@ typedef struct Bytes {
 #define IO_DONE 0
 #define IO_CALL 1
 
+// Result Tags
+#define RESULT_OK  0
+#define RESULT_ERR 1
+
+// IOError = {
+//   Type,           -- a type error
+//   Name,           -- invalid io func name
+//   Inner {val: T}, -- an error while calling an io func
+// }
+#define IO_ERR_TYPE 0
+#define IO_ERR_NAME 1
+#define IO_ERR_INNER 2
+
+typedef struct IOError {
+  u32 tag;
+  Port val;
+} IOError;
+
 // List Type
 #define LIST_NIL  0
 #define LIST_CONS 1
@@ -248,22 +268,225 @@ extern "C" Port gnet_inject_bytes(GNet* gnet, Bytes *bytes) {
   Port* d_ret;
   cudaMalloc(&d_ret, sizeof(Port));
 
-  Bytes cu_bytes;
-  cu_bytes.len = bytes->len;
+  Bytes bytes_cu;
+  bytes_cu.len = bytes->len;
+
+  cudaMalloc(&bytes_cu.buf, sizeof(char) * bytes_cu.len);
+  cudaMemcpy(bytes_cu.buf, bytes->buf, sizeof(char) * bytes_cu.len, cudaMemcpyHostToDevice);
+
+  make_bytes_port<<<1,1>>>(gnet, bytes_cu, d_ret);
+
+  Port ret;
+  cudaMemcpy(&ret, d_ret, sizeof(Port), cudaMemcpyDeviceToHost);
+  cudaFree(d_ret);
+  cudaFree(bytes_cu.buf);
+
+  return ret;
+}
+
+/// Returns a λ-Encoded Ctr for a RESULT_OK: λt ((t RESULT_OK) val)
+__device__ Port inject_ok(Net* net, TM* tm, Port val) {
+  if (!get_resources(net, tm, 0, 3, 1)) {
+    printf("inject_ok: failed to get resources\n");
+    return new_port(ERA, 0);
+  }
+
+  u32 v1 = tm->vloc[0];
+
+  u32 n1 = tm->nloc[0];
+  u32 n2 = tm->nloc[1];
+  u32 n3 = tm->nloc[2];
+
+  vars_create(net, v1, NONE);
+  Port var = new_port(VAR, v1);
+
+  node_create(net, n1, new_pair(val, var));
+  node_create(net, n2, new_pair(new_port(NUM, new_u24(RESULT_OK)), new_port(CON, n1)));
+  node_create(net, n3, new_pair(new_port(CON, n2), var));
+
+  return new_port(CON, n3);
+}
+
+__global__ void make_ok_port(GNet* gnet, Port val, Port* ret) {
+  if (GID() == 0) {
+    TM tm = tmem_new();
+    Net net = vnet_new(gnet, NULL, gnet->turn);
+    *ret = inject_ok(&net, &tm, val);
+  }
+}
+
+extern "C" Port gnet_inject_ok(GNet* gnet, Port val) {
+  Port* d_ret;
+  cudaMalloc(&d_ret, sizeof(Port));
+
+  make_ok_port<<<1,1>>>(gnet, val, d_ret);
+
+  Port ret;
+  cudaMemcpy(&ret, d_ret, sizeof(Port), cudaMemcpyDeviceToHost);
+  cudaFree(d_ret);
+
+  return ret;
+}
+
+/// Returns a λ-Encoded Ctr for a RESULT_ERR: λt ((t RESULT_ERR) err)
+__device__ Port inject_err(Net* net, TM* tm, Port err) {
+  if (!get_resources(net, tm, 0, 3, 1)) {
+    printf("inject_err: failed to get resources\n");
+    return new_port(ERA, 0);
+  }
+
+  u32 v1 = tm->vloc[0];
+
+  u32 n1 = tm->nloc[0];
+  u32 n2 = tm->nloc[1];
+  u32 n3 = tm->nloc[2];
+
+  vars_create(net, v1, NONE);
+  Port var = new_port(VAR, v1);
+
+  node_create(net, n1, new_pair(err, var));
+  node_create(net, n2, new_pair(new_port(NUM, new_u24(RESULT_ERR)), new_port(CON, n1)));
+  node_create(net, n3, new_pair(new_port(CON, n2), var));
+
+  return new_port(CON, n3);
+}
+
+
+__global__ void make_err_port(GNet* gnet, Port val, Port* ret) {
+  if (GID() == 0) {
+    TM tm = tmem_new();
+    Net net = vnet_new(gnet, NULL, gnet->turn);
+    *ret = inject_err(&net, &tm, val);
+  }
+}
+
+extern "C" Port gnet_inject_err(GNet* gnet, Port val) {
+  Port* d_ret;
+  cudaMalloc(&d_ret, sizeof(Port));
+
+  make_err_port<<<1,1>>>(gnet, val, d_ret);
+
+  Port ret;
+  cudaMemcpy(&ret, d_ret, sizeof(Port), cudaMemcpyDeviceToHost);
+  cudaFree(d_ret);
+
+  return ret;
+}
+
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError(..))
+__device__ Port inject_io_err(Net* net, TM* tm, IOError err) {
+  if (err.tag <= IO_ERR_NAME) {
+    if (!get_resources(net, tm, 0, 2, 1)) {
+      return new_port(ERA, 0);
+    }
+
+    u32 v1 = tm->vloc[0];
+
+    u32 n1 = tm->nloc[0];
+    u32 n2 = tm->nloc[1];
+
+    vars_create(net, v1, NONE);
+    Port var = new_port(VAR, v1);
+
+    node_create(net, n1, new_pair(new_port(NUM, new_u24(err.tag)), var));
+    node_create(net, n2, new_pair(new_port(CON, n1), var));
+
+    return inject_err(net, tm, new_port(CON, n2));
+  }
+
+  if (!get_resources(net, tm, 0, 3, 1)) {
+    return new_port(ERA, 0);
+  }
+
+  u32 v1 = tm->vloc[0];
+
+  u32 n1 = tm->nloc[0];
+  u32 n2 = tm->nloc[1];
+  u32 n3 = tm->nloc[2];
+
+  vars_create(net, v1, NONE);
+  Port var = new_port(VAR, v1);
+
+  node_create(net, n1, new_pair(err.val, var));
+  node_create(net, n2, new_pair(new_port(NUM, new_u24(IO_ERR_INNER)), new_port(CON, n1)));
+  node_create(net, n3, new_pair(new_port(CON, n2), var));
 
-  cudaMalloc(&cu_bytes.buf, sizeof(char) * cu_bytes.len);
-  cudaMemcpy(cu_bytes.buf, bytes->buf, sizeof(char) * cu_bytes.len, cudaMemcpyHostToDevice);
+  return inject_err(net, tm, new_port(CON, n3));
+}
 
-  make_bytes_port<<<1,1>>>(gnet, cu_bytes, d_ret);
+__global__ void make_io_err_port(GNet* gnet, IOError err, Port* ret) {
+  if (GID() == 0) {
+    TM tm = tmem_new();
+    Net net = vnet_new(gnet, NULL, gnet->turn);
+    *ret = inject_io_err(&net, &tm, err);
+  }
+}
+
+extern "C" Port gnet_inject_io_err(GNet* gnet, IOError err) {
+  Port* d_ret;
+  cudaMalloc(&d_ret, sizeof(Port));
+
+  make_io_err_port<<<1,1>>>(gnet, err, d_ret);
 
   Port ret;
   cudaMemcpy(&ret, d_ret, sizeof(Port), cudaMemcpyDeviceToHost);
   cudaFree(d_ret);
-  cudaFree(cu_bytes.buf);
 
   return ret;
 }
 
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError/Type)
+extern "C" Port gnet_inject_io_err_type(GNet* gnet) {
+  IOError io_error = {
+    .tag = IO_ERR_TYPE,
+  };
+
+  return gnet_inject_io_err(gnet, io_error);
+}
+
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError/Name)
+extern "C" Port gnet_inject_io_err_name(GNet* gnet) {
+  IOError io_error = {
+    .tag = IO_ERR_NAME,
+  };
+
+  return gnet_inject_io_err(gnet, io_error);
+}
+
+/// Returns a λ-Encoded Ctr for a Result/Err(IOError/Inner(val))
+extern "C" Port gnet_inject_io_err_inner(GNet* gnet, Port val) {
+  IOError io_error = {
+    .tag = IO_ERR_INNER,
+    .val = val,
+  };
+
+  return gnet_inject_io_err(gnet, io_error);
+}
+
+/// Returns a λ-Encoded Ctr for an Result<T, IOError<String>>
+/// `err` must be `NUL`-terminated.
+extern "C" Port gnet_inject_io_err_str(GNet* gnet, char* err) {
+  Port* d_bytes_port;
+  cudaMalloc(&d_bytes_port, sizeof(Port));
+
+  Bytes bytes_cu;
+  bytes_cu.len = strlen(err);
+
+  cudaMalloc(&bytes_cu.buf, sizeof(char) * bytes_cu.len);
+  cudaMemcpy(bytes_cu.buf, err, sizeof(char) * bytes_cu.len, cudaMemcpyHostToDevice);
+
+  make_bytes_port<<<1,1>>>(gnet, bytes_cu, d_bytes_port);
+
+  Port bytes_port;
+  cudaMemcpy(&bytes_port, d_bytes_port, sizeof(Port), cudaMemcpyDeviceToHost);
+  cudaFree(d_bytes_port);
+
+  cudaFree(bytes_cu.buf);
+
+  return gnet_inject_io_err_inner(gnet, bytes_port);
+}
+
+
 // Primitive IO Fns
 // -----------------
 
@@ -321,19 +544,19 @@ void* readback_dylib(Port port) {
 
 // Reads from a file a specified number of bytes.
 // `argm` is a tuple of (file_descriptor, num_bytes).
+// Returns: Result<Bytes, IOError<i24>>
 Port io_read(GNet* gnet, Port argm) {
   Tup tup = gnet_readback_tup(gnet, argm, 2);
   if (tup.elem_len != 2) {
     fprintf(stderr, "io_read: expected 2-tuple\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_type(gnet);
   }
 
   FILE* fp = readback_file(tup.elem_buf[0]);
   u32 num_bytes = get_u24(get_val(tup.elem_buf[1]));
 
   if (fp == NULL) {
-    fprintf(stderr, "io_read: invalid file descriptor\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(EBADF)));
   }
 
   /// Read a string.
@@ -344,23 +567,24 @@ Port io_read(GNet* gnet, Port argm) {
   if ((bytes.len != num_bytes) && ferror(fp)) {
     fprintf(stderr, "io_read: failed to read\n");
     free(bytes.buf);
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(ferror(fp))));
   }
 
   // Convert it to a port.
   Port ret = gnet_inject_bytes(gnet, &bytes);
   free(bytes.buf);
-  return ret;
+
+  return gnet_inject_ok(gnet, ret);
 }
 
 // Opens a file with the provided mode.
 // `argm` is a tuple (CON node) of the
 // file name and mode as strings.
+// Returns: Result<File, IOError<i24>>
 Port io_open(GNet* gnet, Port argm) {
   Tup tup = gnet_readback_tup(gnet, argm, 2);
   if (tup.elem_len != 2) {
-    fprintf(stderr, "io_open: expected 2-tuple\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_type(gnet);
   }
 
   Str name = gnet_readback_str(gnet, tup.elem_buf[0]);
@@ -373,78 +597,81 @@ Port io_open(GNet* gnet, Port argm) {
       free(name.buf);
       free(mode.buf);
 
-      return new_port(NUM, new_u24(fd));
+      if (FILE_POINTERS[fd] == NULL) {
+        return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(errno)));
+      }
+
+      return gnet_inject_ok(gnet, new_port(NUM, new_u24(fd)));
     }
   }
 
-  fprintf(stderr, "io_open: too many open files\n");
-
   free(name.buf);
   free(mode.buf);
 
-  return new_port(ERA, 0);
+  // too many open files
+  return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(EMFILE)));
 }
 
 // Closes a file, reclaiming the file descriptor.
+// Returns: Result<*, IOError<i24>>
 Port io_close(GNet* gnet, Port argm) {
   FILE* fp = readback_file(argm);
   if (fp == NULL) {
-    fprintf(stderr, "io_close: invalid file descriptor\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(EBADF)));
   }
 
-  int err = fclose(fp) != 0;
-  if (err != 0) {
-    fprintf(stderr, "io_close: failed to close: %i\n", err);
-    return new_port(ERA, 0);
+  if (fclose(fp) != 0) {
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(ferror(fp))));
   }
 
   FILE_POINTERS[get_u24(get_val(argm))] = NULL;
-  return new_port(ERA, 0);
-}
-
-// Flushes an output stream.
-Port io_flush(GNet* gnet, Port argm) {
-  FILE* fp = readback_file(argm);
-  if (fp == NULL) {
-    fprintf(stderr, "io_flush: invalid file descriptor\n");
-    return new_port(ERA, 0);
-  }
 
-  int err = fflush(fp) != 0;
-  if (err != 0) {
-    fprintf(stderr, "io_flush: failed to flush: %i\n", err);
-    return new_port(ERA, 0);
-  }
-
-  return new_port(ERA, 0);
+  return gnet_inject_ok(gnet, new_port(ERA, 0));
 }
 
 // Writes a list of bytes to a file.
 // `argm` is a tuple (CON node) of the
 // file descriptor and list of bytes to write.
+// Returns: Result<*, IOError<i24>>
 Port io_write(GNet* gnet, Port argm) {
   Tup tup = gnet_readback_tup(gnet, argm, 2);
   if (tup.elem_len != 2) {
-    fprintf(stderr, "io_write: expected 2-tuple\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_type(gnet);
   }
 
   FILE* fp = readback_file(tup.elem_buf[0]);
   Bytes bytes = gnet_readback_bytes(gnet, tup.elem_buf[1]);
 
   if (fp == NULL) {
-    fprintf(stderr, "io_write: invalid file descriptor\n");
     free(bytes.buf);
-    return new_port(ERA, 0);
+
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(EBADF)));
   }
 
   if (fwrite(bytes.buf, sizeof(char), bytes.len, fp) != bytes.len) {
-    fprintf(stderr, "io_write: failed to write\n");
+    free(bytes.buf);
+
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(ferror(fp))));
   }
 
   free(bytes.buf);
-  return new_port(ERA, 0);
+
+  return gnet_inject_ok(gnet, new_port(ERA, 0));
+}
+
+// Flushes an output stream.
+// Returns: Result<*, IOError<i24>>
+Port io_flush(GNet* gnet, Port argm) {
+  FILE* fp = readback_file(argm);
+  if (fp == NULL) {
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(EBADF)));
+  }
+
+  if (fflush(fp) != 0) {
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(ferror(fp))));
+  }
+
+  return gnet_inject_ok(gnet, new_port(ERA, 0));
 }
 
 // Seeks to a position in a file.
@@ -455,11 +682,12 @@ Port io_write(GNet* gnet, Port argm) {
 //    - 0 (SEEK_SET): beginning of file
 //    - 1 (SEEK_CUR): current position of the file pointer
 //    - 2 (SEEK_END): end of the file
+// Returns: Result<*, IOError<i24>>
 Port io_seek(GNet* gnet, Port argm) {
   Tup tup = gnet_readback_tup(gnet, argm, 3);
   if (tup.elem_len != 3) {
     fprintf(stderr, "io_seek: expected 3-tuple\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_type(gnet);
   }
 
   FILE* fp = readback_file(tup.elem_buf[0]);
@@ -467,8 +695,7 @@ Port io_seek(GNet* gnet, Port argm) {
   u32 whence = get_i24(get_val(tup.elem_buf[2]));
 
   if (fp == NULL) {
-    fprintf(stderr, "io_write: invalid file descriptor\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(EBADF)));
   }
 
   int cwhence;
@@ -477,19 +704,19 @@ Port io_seek(GNet* gnet, Port argm) {
     case 1: cwhence = SEEK_CUR; break;
     case 2: cwhence = SEEK_END; break;
     default:
-      fprintf(stderr, "io_seek: invalid whence\n");
-      return new_port(ERA, 0);
+      return gnet_inject_io_err_type(gnet);
   }
 
   if (fseek(fp, offset, cwhence) != 0) {
-    fprintf(stderr, "io_seek: failed to seek\n");
+    return gnet_inject_io_err_inner(gnet, new_port(NUM, new_i24(ferror(fp))));
   }
 
-  return new_port(ERA, 0);
+  return gnet_inject_ok(gnet, new_port(ERA, 0));
 }
 
 // Returns the current time as a tuple of the high
 // and low 24 bits of a 48-bit nanosecond timestamp.
+// Returns: Result<(u24, u24), IOError<*>>
 Port io_get_time(GNet* gnet, Port argm) {
   // Get the current time in nanoseconds
   u64 time_ns = time64();
@@ -503,11 +730,11 @@ Port io_get_time(GNet* gnet, Port argm) {
 // Sleeps.
 // `argm` is a tuple (CON node) of the high and low
 // 24 bits for a 48-bit duration in nanoseconds.
+// Returns: Result<*, IOError<*>>
 Port io_sleep(GNet* gnet, Port argm) {
   Tup tup = gnet_readback_tup(gnet, argm, 2);
   if (tup.elem_len != 2) {
-    fprintf(stderr, "io_sleep: expected 3-tuple\n");
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_type(gnet);
   }
 
   // Get the sleep duration node
@@ -522,14 +749,15 @@ Port io_sleep(GNet* gnet, Port argm) {
   ts.tv_sec = dur_ns / 1000000000;
   ts.tv_nsec = dur_ns % 1000000000;
   nanosleep(&ts, NULL);
-  // Return an eraser
-  return new_port(ERA, 0);
+
+  return gnet_inject_ok(gnet, new_port(ERA, 0));
 }
 
 // Opens a dylib at the provided path.
 // `argm` is a tuple of `filename` and `lazy`.
 // `filename` is a λ-encoded string.
 // `lazy` is a `bool` indicating if functions should be lazily loaded.
+// Returns: Result<Dylib, IOError<String>>
 Port io_dl_open(GNet* gnet, Port argm) {
   Tup tup = gnet_readback_tup(gnet, argm, 2);
   Str str = gnet_readback_str(gnet, tup.elem_buf[0]);
@@ -540,20 +768,18 @@ Port io_dl_open(GNet* gnet, Port argm) {
   for (u32 dl = 0; dl < sizeof(DYLIBS); dl++) {
     if (DYLIBS[dl] == NULL) {
       DYLIBS[dl] = dlopen(str.buf, flags);
-      if (DYLIBS[dl] == NULL) {
-        fprintf(stderr, "failed to open dylib '%s': %s\n", str.buf, dlerror());
 
-        return new_port(ERA, 0);
-      } else {
-        fprintf(stderr, "opened dylib '%s'\n", str.buf);
+      free(str.buf);
+
+      if (DYLIBS[dl] == NULL) {
+        return gnet_inject_io_err_str(gnet, dlerror());
       }
 
-      return new_port(NUM, new_u24(dl));
+      return gnet_inject_ok(gnet, new_port(NUM, new_u24(dl)));
     }
   }
 
-  fprintf(stderr, "io_dl_open: too many open dylibs\n");
-  return new_port(ERA, 0);
+  return gnet_inject_io_err_str(gnet, "too many open dylibs");
 }
 
 // Calls a function from a loaded dylib.
@@ -561,11 +787,17 @@ Port io_dl_open(GNet* gnet, Port argm) {
 // `dylib_handle` is the numeric node returned from a `DL_OPEN` call.
 // `symbol` is a λ-encoded string of the symbol name.
 // `args` is the argument to be provided to the dylib symbol.
+//
+// This function returns a Result with an Ok variant containing an
+// arbitrary type.
+//
+// Returns Result<T, IOError<String>>
 Port io_dl_call(GNet* gnet, Port argm) {
   Tup tup = gnet_readback_tup(gnet, argm, 3);
   if (tup.elem_len != 3) {
     fprintf(stderr, "io_dl_call: expected 3-tuple\n");
-    return new_port(ERA, 0);
+
+    return gnet_inject_io_err_type(gnet);
   }
 
   void* dl = readback_dylib(tup.elem_buf[0]);
@@ -575,28 +807,31 @@ Port io_dl_call(GNet* gnet, Port argm) {
   Port (*func)(GNet*, Port) = (Port (*)(GNet*, Port)) dlsym(dl, symbol.buf);
   char* error = dlerror();
   if (error != NULL) {
-    fprintf(stderr, "io_dl_call: failed to get symbol '%s': %s\n", symbol.buf, error);
+    return gnet_inject_io_err_str(gnet, error);
   }
 
-  return func(gnet, tup.elem_buf[2]);
+  return gnet_inject_ok(gnet, func(gnet, tup.elem_buf[2]));
 }
 
 // Closes a loaded dylib, reclaiming the handle.
-Port io_dl_close(Net* net, Book* book, Port argm) {
+//
+// Returns:  Result<*, IOError<String>>
+Port io_dl_close(GNet* gnet, Book* book, Port argm) {
   void* dl = readback_dylib(argm);
   if (dl == NULL) {
     fprintf(stderr, "io_dl_close: invalid handle\n");
-    return new_port(ERA, 0);
+
+    return gnet_inject_io_err_type(gnet);
   }
 
   int err = dlclose(dl) != 0;
   if (err != 0) {
-    fprintf(stderr, "io_dl_close: failed to close: %i\n", err);
-    return new_port(ERA, 0);
+    return gnet_inject_io_err_str(gnet, dlerror());
   }
 
   DYLIBS[get_u24(get_val(argm))] = NULL;
-  return new_port(ERA, 0);
+
+  return gnet_inject_ok(gnet, new_port(ERA, 0));
 }
 
 void book_init(Book* book) {
@@ -634,7 +869,7 @@ void do_run_io(GNet* gnet, Book* book, Port port) {
     Ctr ctr = gnet_readback_ctr(gnet, gnet_peek(gnet, port));
 
     // Checks if IO Magic Number is a CON
-    if (get_tag(ctr.args_buf[0]) != CON) {
+    if (ctr.args_len < 1 || get_tag(ctr.args_buf[0]) != CON) {
       break;
     }
 
@@ -647,6 +882,11 @@ void do_run_io(GNet* gnet, Book* book, Port port) {
 
     switch (ctr.tag) {
       case IO_CALL: {
+        if (ctr.args_len != 4) {
+          fprintf(stderr, "invalid IO_CALL: args_len = %u\n", ctr.args_len);
+          break;
+        }
+
         Str  func = gnet_readback_str(gnet, ctr.args_buf[1]);
         FFn* ffn  = NULL;
         // FIXME: optimize this linear search
@@ -657,27 +897,25 @@ void do_run_io(GNet* gnet, Book* book, Port port) {
           }
         }
 
-        if (ffn == NULL) {
-          fprintf(stderr, "Unknown IO func '%s'\n", func.buf);
-
-          free(func.buf);
-
-          break;
-        }
-
-        debug("running io func '%s'\n", func.buf);
-
         free(func.buf);
 
         Port argm = ctr.args_buf[2];
         Port cont = ctr.args_buf[3];
-        Port ret  = ffn->func(gnet, argm);
+
+        Port ret;
+        if (ffn == NULL) {
+          ret = gnet_inject_io_err_name(gnet);
+        } else {
+          ret = ffn->func(gnet, argm);
+        }
 
         Port p = gnet_make_node(gnet, CON, ret, ROOT);
         gnet_boot_redex(gnet, new_pair(p, cont));
         port = ROOT;
+
         continue;
       }
+
       case IO_DONE: {
         break;
       }
diff --git a/tests/programs/io/basic.bend b/tests/programs/io/basic.bend
new file mode 100644
index 00000000..bce1eada
--- /dev/null
+++ b/tests/programs/io/basic.bend
@@ -0,0 +1,35 @@
+test-io = 1
+
+def unwrap(res):
+  match res:
+    case Result/Ok:
+      return res.val
+    case Result/Err:
+      return res.val
+
+def open():
+  return call("OPEN", ("./LICENSE", "r"))
+
+def read(f):
+  return call("READ", (f, 47))
+
+def print(bytes):
+  with IO:
+    * <- call("WRITE", (1, bytes))
+    * <- call("WRITE", (1, "\n"))
+
+    return wrap(*)
+
+def close(f):
+  return call("CLOSE", f)
+
+def main():
+  with IO:
+    f <- open()
+    f = unwrap(f)
+    bytes <- read(f)
+    bytes = unwrap(bytes)
+    * <- print(bytes)
+    res <- close(f)
+
+    return wrap(res)
diff --git a/tests/programs/io/basic.hvm b/tests/programs/io/basic.hvm
new file mode 100644
index 00000000..c4b72892
--- /dev/null
+++ b/tests/programs/io/basic.hvm
@@ -0,0 +1,115 @@
+@IO/Call = (a (b (c (d ((@IO/Call/tag (a (b (c (d e))))) e)))))
+
+@IO/Call/tag = 1
+
+@IO/Done = (a (b ((@IO/Done/tag (a (b c))) c)))
+
+@IO/Done/tag = 0
+
+@IO/MAGIC = (13683217 16719857)
+
+@IO/bind = ((@IO/bind__C2 a) a)
+
+@IO/bind__C0 = (* (b (a c)))
+  & @undefer ~ (a (b c))
+
+@IO/bind__C1 = (* (* (a (b ((c d) (e g))))))
+  & @IO/Call ~ (@IO/MAGIC (a (b ((c f) g))))
+  & @IO/bind ~ (d (e f))
+
+@IO/bind__C2 = (?((@IO/bind__C0 @IO/bind__C1) a) a)
+
+@IO/wrap = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@String/Cons = (a (b ((@String/Cons/tag (a (b c))) c)))
+
+@String/Cons/tag = 1
+
+@String/Nil = ((@String/Nil/tag a) a)
+
+@String/Nil/tag = 0
+
+@call = (a (b c))
+  & @IO/Call ~ (@IO/MAGIC (a (b (@call__C0 c))))
+
+@call__C0 = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@close = f
+  & @call ~ (e f)
+  & @String/Cons ~ (67 (d e))
+  & @String/Cons ~ (76 (c d))
+  & @String/Cons ~ (79 (b c))
+  & @String/Cons ~ (83 (a b))
+  & @String/Cons ~ (69 (@String/Nil a))
+
+@main = w
+  & @IO/bind ~ (@open ((((s (a u)) (@IO/wrap v)) v) w))
+  & @IO/bind ~ (c ((((n (o (d q))) (r (s t))) t) u))
+  & @unwrap ~ (a {b r})
+  & @read ~ (b c)
+  & @IO/bind ~ (f ((((g (k (* m))) (n (o p))) p) q))
+  & @print ~ (e f)
+  & @unwrap ~ (d e)
+  & @IO/bind ~ (h ((((i i) (k l)) l) m))
+  & @close ~ (g h)
+
+@open = o
+  & @call ~ (d ((m n) o))
+  & @String/Cons ~ (79 (c d))
+  & @String/Cons ~ (80 (b c))
+  & @String/Cons ~ (69 (a b))
+  & @String/Cons ~ (78 (@String/Nil a))
+  & @String/Cons ~ (46 (l m))
+  & @String/Cons ~ (47 (k l))
+  & @String/Cons ~ (76 (j k))
+  & @String/Cons ~ (73 (i j))
+  & @String/Cons ~ (67 (h i))
+  & @String/Cons ~ (69 (g h))
+  & @String/Cons ~ (78 (f g))
+  & @String/Cons ~ (83 (e f))
+  & @String/Cons ~ (69 (@String/Nil e))
+  & @String/Cons ~ (114 (@String/Nil n))
+
+@print = (f h)
+  & @IO/bind ~ (g (@print__C3 h))
+  & @call ~ (e ((1 f) g))
+  & @String/Cons ~ (87 (d e))
+  & @String/Cons ~ (82 (c d))
+  & @String/Cons ~ (73 (b c))
+  & @String/Cons ~ (84 (a b))
+  & @String/Cons ~ (69 (@String/Nil a))
+
+@print__C0 = ((* a) (* a))
+
+@print__C1 = g
+  & @call ~ (e ((1 f) g))
+  & @String/Cons ~ (87 (d e))
+  & @String/Cons ~ (82 (c d))
+  & @String/Cons ~ (73 (b c))
+  & @String/Cons ~ (84 (a b))
+  & @String/Cons ~ (69 (@String/Nil a))
+  & @String/Cons ~ (10 (@String/Nil f))
+
+@print__C2 = (a (* c))
+  & @IO/bind ~ (@print__C1 (((@print__C0 (a b)) b) c))
+
+@print__C3 = ((@print__C2 (@IO/wrap a)) a)
+
+@read = (e f)
+  & @call ~ (d ((e 47) f))
+  & @String/Cons ~ (82 (c d))
+  & @String/Cons ~ (69 (b c))
+  & @String/Cons ~ (65 (a b))
+  & @String/Cons ~ (68 (@String/Nil a))
+
+@test-io = 1
+
+@undefer = (((a a) b) b)
+
+@unwrap = ((@unwrap__C0 a) a)
+
+@unwrap__C0 = (?(((a a) (* (b b))) c) c)
+
+
diff --git a/tests/programs/io/invalid-name.bend b/tests/programs/io/invalid-name.bend
new file mode 100644
index 00000000..6bc96c84
--- /dev/null
+++ b/tests/programs/io/invalid-name.bend
@@ -0,0 +1,7 @@
+test-io = 1
+
+def main():
+  with IO:
+    f <- call("INVALID-NAME", ("./README.md", "r"))
+
+    return wrap(f)
diff --git a/tests/programs/io/invalid-name.hvm b/tests/programs/io/invalid-name.hvm
new file mode 100644
index 00000000..e73aa507
--- /dev/null
+++ b/tests/programs/io/invalid-name.hvm
@@ -0,0 +1,71 @@
+@IO/Call = (a (b (c (d ((@IO/Call/tag (a (b (c (d e))))) e)))))
+
+@IO/Call/tag = 1
+
+@IO/Done = (a (b ((@IO/Done/tag (a (b c))) c)))
+
+@IO/Done/tag = 0
+
+@IO/MAGIC = (13683217 16719857)
+
+@IO/bind = ((@IO/bind__C2 a) a)
+
+@IO/bind__C0 = (* (b (a c)))
+  & @undefer ~ (a (b c))
+
+@IO/bind__C1 = (* (* (a (b ((c d) (e g))))))
+  & @IO/Call ~ (@IO/MAGIC (a (b ((c f) g))))
+  & @IO/bind ~ (d (e f))
+
+@IO/bind__C2 = (?((@IO/bind__C0 @IO/bind__C1) a) a)
+
+@IO/wrap = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@String/Cons = (a (b ((@String/Cons/tag (a (b c))) c)))
+
+@String/Cons/tag = 1
+
+@String/Nil = ((@String/Nil/tag a) a)
+
+@String/Nil/tag = 0
+
+@call = (a (b c))
+  & @IO/Call ~ (@IO/MAGIC (a (b (@call__C0 c))))
+
+@call__C0 = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@main = cb
+  & @IO/bind ~ (y ((((z z) (@IO/wrap bb)) bb) cb))
+  & @call ~ (l ((w x) y))
+  & @String/Cons ~ (73 (k l))
+  & @String/Cons ~ (78 (j k))
+  & @String/Cons ~ (86 (i j))
+  & @String/Cons ~ (65 (h i))
+  & @String/Cons ~ (76 (g h))
+  & @String/Cons ~ (73 (f g))
+  & @String/Cons ~ (68 (e f))
+  & @String/Cons ~ (45 (d e))
+  & @String/Cons ~ (78 (c d))
+  & @String/Cons ~ (65 (b c))
+  & @String/Cons ~ (77 (a b))
+  & @String/Cons ~ (69 (@String/Nil a))
+  & @String/Cons ~ (46 (v w))
+  & @String/Cons ~ (47 (u v))
+  & @String/Cons ~ (82 (t u))
+  & @String/Cons ~ (69 (s t))
+  & @String/Cons ~ (65 (r s))
+  & @String/Cons ~ (68 (q r))
+  & @String/Cons ~ (77 (p q))
+  & @String/Cons ~ (69 (o p))
+  & @String/Cons ~ (46 (n o))
+  & @String/Cons ~ (109 (m n))
+  & @String/Cons ~ (100 (@String/Nil m))
+  & @String/Cons ~ (114 (@String/Nil x))
+
+@test-io = 1
+
+@undefer = (((a a) b) b)
+
+
diff --git a/tests/programs/io/open1.bend b/tests/programs/io/open1.bend
new file mode 100644
index 00000000..3353c50d
--- /dev/null
+++ b/tests/programs/io/open1.bend
@@ -0,0 +1,7 @@
+test-io = 1
+
+def main():
+  with IO:
+    f <- call("OPEN", ("./LICENSE", "r"))
+
+    return wrap(f)
diff --git a/tests/programs/io/open1.hvm b/tests/programs/io/open1.hvm
new file mode 100644
index 00000000..2b433d47
--- /dev/null
+++ b/tests/programs/io/open1.hvm
@@ -0,0 +1,61 @@
+@IO/Call = (a (b (c (d ((@IO/Call/tag (a (b (c (d e))))) e)))))
+
+@IO/Call/tag = 1
+
+@IO/Done = (a (b ((@IO/Done/tag (a (b c))) c)))
+
+@IO/Done/tag = 0
+
+@IO/MAGIC = (13683217 16719857)
+
+@IO/bind = ((@IO/bind__C2 a) a)
+
+@IO/bind__C0 = (* (b (a c)))
+  & @undefer ~ (a (b c))
+
+@IO/bind__C1 = (* (* (a (b ((c d) (e g))))))
+  & @IO/Call ~ (@IO/MAGIC (a (b ((c f) g))))
+  & @IO/bind ~ (d (e f))
+
+@IO/bind__C2 = (?((@IO/bind__C0 @IO/bind__C1) a) a)
+
+@IO/wrap = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@String/Cons = (a (b ((@String/Cons/tag (a (b c))) c)))
+
+@String/Cons/tag = 1
+
+@String/Nil = ((@String/Nil/tag a) a)
+
+@String/Nil/tag = 0
+
+@call = (a (b c))
+  & @IO/Call ~ (@IO/MAGIC (a (b (@call__C0 c))))
+
+@call__C0 = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@main = s
+  & @IO/bind ~ (o ((((p p) (@IO/wrap r)) r) s))
+  & @call ~ (d ((m n) o))
+  & @String/Cons ~ (79 (c d))
+  & @String/Cons ~ (80 (b c))
+  & @String/Cons ~ (69 (a b))
+  & @String/Cons ~ (78 (@String/Nil a))
+  & @String/Cons ~ (46 (l m))
+  & @String/Cons ~ (47 (k l))
+  & @String/Cons ~ (76 (j k))
+  & @String/Cons ~ (73 (i j))
+  & @String/Cons ~ (67 (h i))
+  & @String/Cons ~ (69 (g h))
+  & @String/Cons ~ (78 (f g))
+  & @String/Cons ~ (83 (e f))
+  & @String/Cons ~ (69 (@String/Nil e))
+  & @String/Cons ~ (114 (@String/Nil n))
+
+@test-io = 1
+
+@undefer = (((a a) b) b)
+
+
diff --git a/tests/programs/io/open2.bend b/tests/programs/io/open2.bend
new file mode 100644
index 00000000..2d06ae9e
--- /dev/null
+++ b/tests/programs/io/open2.bend
@@ -0,0 +1,7 @@
+test-io = 1
+
+def main():
+  with IO:
+    f <- call("OPEN", ("fake-file", "r"))
+
+    return wrap(f)
diff --git a/tests/programs/io/open2.hvm b/tests/programs/io/open2.hvm
new file mode 100644
index 00000000..9c790cda
--- /dev/null
+++ b/tests/programs/io/open2.hvm
@@ -0,0 +1,61 @@
+@IO/Call = (a (b (c (d ((@IO/Call/tag (a (b (c (d e))))) e)))))
+
+@IO/Call/tag = 1
+
+@IO/Done = (a (b ((@IO/Done/tag (a (b c))) c)))
+
+@IO/Done/tag = 0
+
+@IO/MAGIC = (13683217 16719857)
+
+@IO/bind = ((@IO/bind__C2 a) a)
+
+@IO/bind__C0 = (* (b (a c)))
+  & @undefer ~ (a (b c))
+
+@IO/bind__C1 = (* (* (a (b ((c d) (e g))))))
+  & @IO/Call ~ (@IO/MAGIC (a (b ((c f) g))))
+  & @IO/bind ~ (d (e f))
+
+@IO/bind__C2 = (?((@IO/bind__C0 @IO/bind__C1) a) a)
+
+@IO/wrap = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@String/Cons = (a (b ((@String/Cons/tag (a (b c))) c)))
+
+@String/Cons/tag = 1
+
+@String/Nil = ((@String/Nil/tag a) a)
+
+@String/Nil/tag = 0
+
+@call = (a (b c))
+  & @IO/Call ~ (@IO/MAGIC (a (b (@call__C0 c))))
+
+@call__C0 = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@main = s
+  & @IO/bind ~ (o ((((p p) (@IO/wrap r)) r) s))
+  & @call ~ (d ((m n) o))
+  & @String/Cons ~ (79 (c d))
+  & @String/Cons ~ (80 (b c))
+  & @String/Cons ~ (69 (a b))
+  & @String/Cons ~ (78 (@String/Nil a))
+  & @String/Cons ~ (102 (l m))
+  & @String/Cons ~ (97 (k l))
+  & @String/Cons ~ (107 (j k))
+  & @String/Cons ~ (101 (i j))
+  & @String/Cons ~ (45 (h i))
+  & @String/Cons ~ (102 (g h))
+  & @String/Cons ~ (105 (f g))
+  & @String/Cons ~ (108 (e f))
+  & @String/Cons ~ (101 (@String/Nil e))
+  & @String/Cons ~ (114 (@String/Nil n))
+
+@test-io = 1
+
+@undefer = (((a a) b) b)
+
+
diff --git a/tests/programs/io/open3.bend b/tests/programs/io/open3.bend
new file mode 100644
index 00000000..518b98da
--- /dev/null
+++ b/tests/programs/io/open3.bend
@@ -0,0 +1,8 @@
+test-io = 1
+
+def main():
+  with IO:
+    # calling open with an unexpected type of arg
+    f <- call("OPEN", 123)
+
+    return wrap(f)
diff --git a/tests/programs/io/open3.hvm b/tests/programs/io/open3.hvm
new file mode 100644
index 00000000..d6e02fce
--- /dev/null
+++ b/tests/programs/io/open3.hvm
@@ -0,0 +1,51 @@
+@IO/Call = (a (b (c (d ((@IO/Call/tag (a (b (c (d e))))) e)))))
+
+@IO/Call/tag = 1
+
+@IO/Done = (a (b ((@IO/Done/tag (a (b c))) c)))
+
+@IO/Done/tag = 0
+
+@IO/MAGIC = (13683217 16719857)
+
+@IO/bind = ((@IO/bind__C2 a) a)
+
+@IO/bind__C0 = (* (b (a c)))
+  & @undefer ~ (a (b c))
+
+@IO/bind__C1 = (* (* (a (b ((c d) (e g))))))
+  & @IO/Call ~ (@IO/MAGIC (a (b ((c f) g))))
+  & @IO/bind ~ (d (e f))
+
+@IO/bind__C2 = (?((@IO/bind__C0 @IO/bind__C1) a) a)
+
+@IO/wrap = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@String/Cons = (a (b ((@String/Cons/tag (a (b c))) c)))
+
+@String/Cons/tag = 1
+
+@String/Nil = ((@String/Nil/tag a) a)
+
+@String/Nil/tag = 0
+
+@call = (a (b c))
+  & @IO/Call ~ (@IO/MAGIC (a (b (@call__C0 c))))
+
+@call__C0 = a
+  & @IO/Done ~ (@IO/MAGIC a)
+
+@main = i
+  & @IO/bind ~ (e ((((f f) (@IO/wrap h)) h) i))
+  & @call ~ (d (123 e))
+  & @String/Cons ~ (79 (c d))
+  & @String/Cons ~ (80 (b c))
+  & @String/Cons ~ (69 (a b))
+  & @String/Cons ~ (78 (@String/Nil a))
+
+@test-io = 1
+
+@undefer = (((a a) b) b)
+
+
diff --git a/tests/programs/io/read_and_print.hvm b/tests/programs/io/read_and_print.hvm
deleted file mode 100644
index bcdc54cf..00000000
--- a/tests/programs/io/read_and_print.hvm
+++ /dev/null
@@ -1,233 +0,0 @@
-@IO_T/Call = (a (b (c (d ((@IO_T/Call/tag (a (b (c (d e))))) e)))))
-
-@IO_T/Call/tag = 1
-
-@IO_T/Done = (a (b ((@IO_T/Done/tag (a (b c))) c)))
-
-@IO_T/Done/tag = 0
-
-@IO_T/MAGIC = (13683217 16719857)
-
-@IO_T/bind = ((@IO_T/bind__C2 a) a)
-
-@IO_T/bind__C0 = (* (b (a c)))
-  & @undefer ~ (a (b c))
-
-@IO_T/bind__C1 = (* (* (a (b ((c d) (e g))))))
-  & @IO_T/Call ~ (@IO_T/MAGIC (a (b ((c f) g))))
-  & @IO_T/bind ~ (d (e f))
-
-@IO_T/bind__C2 = (?((@IO_T/bind__C0 @IO_T/bind__C1) a) a)
-
-@String/Cons = (a (b ((@String/Cons/tag (a (b c))) c)))
-
-@String/Cons/tag = 1
-
-@String/Nil = ((@String/Nil/tag a) a)
-
-@String/Nil/tag = 0
-
-@call_io = (a (b c))
-  & @IO_T/Call ~ (@IO_T/MAGIC (a (b (@call_io__C0 c))))
-
-@call_io__C0 = a
-  & @IO_T/Done ~ (@IO_T/MAGIC a)
-
-@main = m
-  & @IO_T/bind ~ (@write_to_file ((((* k) l) l) m))
-  & @IO_T/bind ~ (@read_from_file ((((* i) j) j) k))
-  & @IO_T/bind ~ (g ((((* 42) h) h) i))
-  & @call_io ~ (e ((1 f) g))
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-  & @String/Cons ~ (10 (@String/Nil f))
-
-@read_from_file = a
-  & @IO_T/bind ~ (@read_from_file__C11 (@read_from_file__C10 a))
-
-@read_from_file__C0 = (f (* g))
-  & @call_io ~ (e (f g))
-  & @String/Cons ~ (67 (d e))
-  & @String/Cons ~ (76 (c d))
-  & @String/Cons ~ (79 (b c))
-  & @String/Cons ~ (83 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@read_from_file__C1 = e
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@read_from_file__C10 = ((@read_from_file__C9 a) a)
-
-@read_from_file__C11 = q
-  & @call_io ~ (d ((o p) q))
-  & @String/Cons ~ (79 (c d))
-  & @String/Cons ~ (80 (b c))
-  & @String/Cons ~ (69 (a b))
-  & @String/Cons ~ (78 (@String/Nil a))
-  & @String/Cons ~ (116 (n o))
-  & @String/Cons ~ (101 (m n))
-  & @String/Cons ~ (115 (l m))
-  & @String/Cons ~ (116 (k l))
-  & @String/Cons ~ (105 (j k))
-  & @String/Cons ~ (110 (i j))
-  & @String/Cons ~ (103 (h i))
-  & @String/Cons ~ (46 (g h))
-  & @String/Cons ~ (116 (f g))
-  & @String/Cons ~ (120 (e f))
-  & @String/Cons ~ (116 (@String/Nil e))
-  & @String/Cons ~ (114 (@String/Nil p))
-
-@read_from_file__C2 = (c (a e))
-  & @IO_T/bind ~ (b (((@read_from_file__C0 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C1 ((1 a) b))
-
-@read_from_file__C3 = d
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (65 (a b))
-  & @String/Cons ~ (68 (@String/Nil a))
-
-@read_from_file__C4 = ({a c} (* e))
-  & @IO_T/bind ~ (b (((@read_from_file__C2 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C3 ((a 5) b))
-
-@read_from_file__C5 = d
-  & @String/Cons ~ (83 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (69 (a b))
-  & @String/Cons ~ (75 (@String/Nil a))
-
-@read_from_file__C6 = ({a c} (* e))
-  & @IO_T/bind ~ (b (((@read_from_file__C4 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C5 ((a (2 0)) b))
-
-@read_from_file__C7 = e
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@read_from_file__C8 = (c (a e))
-  & @IO_T/bind ~ (b (((@read_from_file__C6 (c d)) d) e))
-  & @call_io ~ (@read_from_file__C7 ((1 a) b))
-
-@read_from_file__C9 = ({e g} i)
-  & @IO_T/bind ~ (f (((@read_from_file__C8 (g h)) h) i))
-  & @call_io ~ (d ((e 5) f))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (65 (a b))
-  & @String/Cons ~ (68 (@String/Nil a))
-
-@read_input = a
-  & @IO_T/bind ~ (@read_input__C2 (@read_input__C1 a))
-
-@read_input__C0 = (* e)
-  & @call_io ~ (d ((0 10) e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (69 (b c))
-  & @String/Cons ~ (65 (a b))
-  & @String/Cons ~ (68 (@String/Nil a))
-
-@read_input__C1 = ((@read_input__C0 a) a)
-
-@read_input__C2 = y
-  & @call_io ~ (e ((1 x) y))
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-  & @String/Cons ~ (87 (w x))
-  & @String/Cons ~ (104 (v w))
-  & @String/Cons ~ (97 (u v))
-  & @String/Cons ~ (116 (t u))
-  & @String/Cons ~ (32 (s t))
-  & @String/Cons ~ (105 (r s))
-  & @String/Cons ~ (115 (q r))
-  & @String/Cons ~ (32 (p q))
-  & @String/Cons ~ (121 (o p))
-  & @String/Cons ~ (111 (n o))
-  & @String/Cons ~ (117 (m n))
-  & @String/Cons ~ (114 (l m))
-  & @String/Cons ~ (32 (k l))
-  & @String/Cons ~ (110 (j k))
-  & @String/Cons ~ (97 (i j))
-  & @String/Cons ~ (109 (h i))
-  & @String/Cons ~ (101 (g h))
-  & @String/Cons ~ (63 (f g))
-  & @String/Cons ~ (10 (@String/Nil f))
-
-@test-io = 1
-
-@undefer = (((a a) b) b)
-
-@write_to_file = a
-  & @IO_T/bind ~ (@write_to_file__C8 (@write_to_file__C7 a))
-
-@write_to_file__C0 = (f (* g))
-  & @call_io ~ (e (f g))
-  & @String/Cons ~ (67 (d e))
-  & @String/Cons ~ (76 (c d))
-  & @String/Cons ~ (79 (b c))
-  & @String/Cons ~ (83 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@write_to_file__C1 = a
-  & @String/Cons ~ (10 (@String/Nil a))
-
-@write_to_file__C2 = e
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@write_to_file__C3 = ({a c} (* e))
-  & @IO_T/bind ~ (b (((@write_to_file__C0 (c d)) d) e))
-  & @call_io ~ (@write_to_file__C2 ((a @write_to_file__C1) b))
-
-@write_to_file__C4 = e
-  & @String/Cons ~ (87 (d e))
-  & @String/Cons ~ (82 (c d))
-  & @String/Cons ~ (73 (b c))
-  & @String/Cons ~ (84 (a b))
-  & @String/Cons ~ (69 (@String/Nil a))
-
-@write_to_file__C5 = ({a d} (b f))
-  & @IO_T/bind ~ (c (((@write_to_file__C3 (d e)) e) f))
-  & @call_io ~ (@write_to_file__C4 ((a b) c))
-
-@write_to_file__C6 = (a c)
-  & @IO_T/bind ~ (@read_input (((@write_to_file__C5 (a b)) b) c))
-
-@write_to_file__C7 = ((@write_to_file__C6 a) a)
-
-@write_to_file__C8 = q
-  & @call_io ~ (d ((o p) q))
-  & @String/Cons ~ (79 (c d))
-  & @String/Cons ~ (80 (b c))
-  & @String/Cons ~ (69 (a b))
-  & @String/Cons ~ (78 (@String/Nil a))
-  & @String/Cons ~ (116 (n o))
-  & @String/Cons ~ (101 (m n))
-  & @String/Cons ~ (115 (l m))
-  & @String/Cons ~ (116 (k l))
-  & @String/Cons ~ (105 (j k))
-  & @String/Cons ~ (110 (i j))
-  & @String/Cons ~ (103 (h i))
-  & @String/Cons ~ (46 (g h))
-  & @String/Cons ~ (116 (f g))
-  & @String/Cons ~ (120 (e f))
-  & @String/Cons ~ (116 (@String/Nil e))
-  & @String/Cons ~ (119 (@String/Nil p))
-
-
diff --git a/tests/snapshots/run__io_file@demo_io__main.hvm.snap b/tests/snapshots/run__io_file@demo_io__main.hvm.snap
index 8503fe52..66b51367 100644
--- a/tests/snapshots/run__io_file@demo_io__main.hvm.snap
+++ b/tests/snapshots/run__io_file@demo_io__main.hvm.snap
@@ -3,6 +3,5 @@ source: tests/run.rs
 expression: c_output
 input_file: examples/demo_io/main.hvm
 ---
-What is your name?
-io fr from
-Result: 42
+                                 Apache License
+Result: ((@IO/Done/tag (@IO/MAGIC (((0 (* x0)) x0) x1))) x1)
diff --git a/tests/snapshots/run__io_file@io__basic.hvm.snap b/tests/snapshots/run__io_file@io__basic.hvm.snap
new file mode 100644
index 00000000..1de699b5
--- /dev/null
+++ b/tests/snapshots/run__io_file@io__basic.hvm.snap
@@ -0,0 +1,7 @@
+---
+source: tests/run.rs
+expression: c_output
+input_file: tests/programs/io/basic.hvm
+---
+                                 Apache License
+Result: ((@IO/Done/tag (@IO/MAGIC (((0 (* x0)) x0) x1))) x1)
diff --git a/tests/snapshots/run__io_file@io__invalid-name.hvm.snap b/tests/snapshots/run__io_file@io__invalid-name.hvm.snap
new file mode 100644
index 00000000..4fa1086c
--- /dev/null
+++ b/tests/snapshots/run__io_file@io__invalid-name.hvm.snap
@@ -0,0 +1,6 @@
+---
+source: tests/run.rs
+expression: c_output
+input_file: tests/programs/io/invalid-name.hvm
+---
+Result: ((@IO/Done/tag (@IO/MAGIC (((1 (((1 x0) x0) x1)) x1) x2))) x2)
diff --git a/tests/snapshots/run__io_file@io__open1.hvm.snap b/tests/snapshots/run__io_file@io__open1.hvm.snap
new file mode 100644
index 00000000..ad0cdc85
--- /dev/null
+++ b/tests/snapshots/run__io_file@io__open1.hvm.snap
@@ -0,0 +1,6 @@
+---
+source: tests/run.rs
+expression: c_output
+input_file: tests/programs/io/open1.hvm
+---
+Result: ((@IO/Done/tag (@IO/MAGIC (((0 (3 x0)) x0) x1))) x1)
diff --git a/tests/snapshots/run__io_file@io__open2.hvm.snap b/tests/snapshots/run__io_file@io__open2.hvm.snap
new file mode 100644
index 00000000..9db2db58
--- /dev/null
+++ b/tests/snapshots/run__io_file@io__open2.hvm.snap
@@ -0,0 +1,6 @@
+---
+source: tests/run.rs
+expression: c_output
+input_file: tests/programs/io/open2.hvm
+---
+Result: ((@IO/Done/tag (@IO/MAGIC (((1 (((2 (+2 x0)) x0) x1)) x1) x2))) x2)
diff --git a/tests/snapshots/run__io_file@io__open3.hvm.snap b/tests/snapshots/run__io_file@io__open3.hvm.snap
new file mode 100644
index 00000000..166b324e
--- /dev/null
+++ b/tests/snapshots/run__io_file@io__open3.hvm.snap
@@ -0,0 +1,6 @@
+---
+source: tests/run.rs
+expression: c_output
+input_file: tests/programs/io/open3.hvm
+---
+Result: ((@IO/Done/tag (@IO/MAGIC (((1 (((0 x0) x0) x1)) x1) x2))) x2)