Fix bugs in implem of Bitmask and Pack

DadaIsCrazy · DadaIsCrazy · commit 3717f533d43e · 2021-06-07T13:27:51.000+02:00
diff --git a/arch/AVX.h b/arch/AVX.h
@@ -63,6 +63,14 @@
 #define LIFT_32(x) _mm256_set1_epi32(x)
 #define LIFT_64(x) _mm256_set1_epi64x(x)
 
+
+#define BITMASK(x,n,c) _mm256_sub_epi##c(ZERO, __mm256_and_si256(_mm256_slli_epi##c(x,n), _mm256_set1_epi##c(1)))
+
+#define PACK_8x2_to_16(a,b)  /* TODO: implement with shuffles */
+#define PACK_16x2_to_32(a,b) /* TODO: implement with shuffles */
+#define PACK_32x2_to_64(a,b) /* TODO: implement with shuffles */
+
+
 #define DATATYPE __m256i
 
 #define SET_ALL_ONE()  ONES
diff --git a/arch/AVX512.h b/arch/AVX512.h
@@ -71,6 +71,13 @@
 #define LIFT_32(x) _mm512_set1_epi32(x)
 #define LIFT_64(x) _mm512_set1_epi64(x)
 
+
+#define BITMASK(x,n,c) _mm512_sub_epi##c(ZERO, __mm512_and_si512(_mm512_slli_epi##c(x,n), _mm512_set1_epi##c(1)))
+
+#define PACK_8x2_to_16(a,b)  /* TODO: implement with shuffles */
+#define PACK_16x2_to_32(a,b) /* TODO: implement with shuffles */
+#define PACK_32x2_to_64(a,b) /* TODO: implement with shuffles */
+
 #define DATATYPE __m512i
 
 #define SET_ALL_ONE()  ONES
diff --git a/arch/SSE.h b/arch/SSE.h
@@ -68,6 +68,11 @@
 #define LIFT_32(x) _mm_set1_epi32(x)
 #define LIFT_64(x) _mm_set1_epi64x(x)
 
+#define BITMASK(x,n,c) _mm_sub_epi##c(ZERO, __mm_and_si128(_mm_slli_epi##c(x,n), _mm_set1_epi##c(1)))
+
+#define PACK_8x2_to_16(a,b)  /* TODO: implement with shuffles */
+#define PACK_16x2_to_32(a,b) /* TODO: implement with shuffles */
+#define PACK_32x2_to_64(a,b) /* TODO: implement with shuffles */
 
 #define ORTHOGONALIZE(in,out) orthogonalize(in,out)
 #define UNORTHOGONALIZE(in,out) unorthogonalize(in,out)
diff --git a/arch/STD.h b/arch/STD.h
@@ -56,6 +56,13 @@
 #define LIFT_32(x) (x)
 #define LIFT_64(x) (x)
 
+#define BITMASK(x,n,c) -(((x) >> (n)) & 1)
+
+#define PACK_8x2_to_16(a,b)  ((((uint16_t)(a)) << 8) | ((uint16_t) (b)))
+#define PACK_16x2_to_32(a,b) ((((uint32_t)(a)) << 16) | ((uint32_t) (b)))
+#define PACK_32x2_to_64(a,b) ((((uint64_t)(a)) << 32) | ((uint64_t) (b)))
+
+
 #define refresh(x,y) *(y) = x
 
 #ifndef DATATYPE
diff --git a/src/c_gen/nodes_to_c.ml b/src/c_gen/nodes_to_c.ml
@@ -118,6 +118,23 @@ let rec expr_to_c (m:mtyp)
                                  (List.length l)
                                  (var_to_c lift_env env v)
                                  (join "," (List.map string_of_int l))
+  | Bitmask(e',ae) -> sprintf "BITMASK(%s, %s, %d)"
+                    (expr_to_c m lift_env env env_var e')
+                    (aexpr_to_c ae)
+                    (match m with
+                     | Mint m_val -> m_val
+                     | _ -> assert false)
+  | Pack(e1,e2,Some typ) ->
+     let args_m = get_type_m (get_normed_expr_type env_var e1) in
+     sprintf "PACK_%dx2_to_%d(%s,%s)"
+       (match args_m with
+        | Mint m_val -> m_val
+        | _ -> assert false)
+       (match m with
+        | Mint m_val -> m_val
+        | _ -> assert false)
+       (expr_to_c args_m lift_env env env_var e1)
+       (expr_to_c args_m lift_env env env_var e2)
   | Shift(op,e,ae) ->
      sprintf "%s(%s,%s,%d)"
              (shift_op_to_c op)
diff --git a/src/normalization/clean.ml b/src/normalization/clean.ml
@@ -12,18 +12,18 @@ let rec collect_var env (var:var) : unit =
 
 let rec collect_expr env (e:expr) : unit =
   match e with
-  | Const _      -> ()
-  | ExpVar(v)    -> collect_var env v
-  | Tuple l      -> List.iter (collect_expr env) l
-  | Not e        -> collect_expr env e
-  | Log(_,x,y)   -> collect_expr env x; collect_expr env y
-  | Arith(_,x,y) -> collect_expr env x; collect_expr env y
-  | Shift(_,x,_) -> collect_expr env x
-  | Shuffle(v,_) -> collect_var env v
-  | Mask(e,_)    -> collect_expr env e
-  | Pack(l,_)    -> List.iter (collect_expr env) l
-  | Fun(_,l)     -> List.iter (collect_expr env) l
-  | Fun_v(_,_,l) -> List.iter (collect_expr env) l
+  | Const _       -> ()
+  | ExpVar(v)     -> collect_var env v
+  | Tuple l       -> List.iter (collect_expr env) l
+  | Not e         -> collect_expr env e
+  | Log(_,x,y)    -> collect_expr env x; collect_expr env y
+  | Arith(_,x,y)  -> collect_expr env x; collect_expr env y
+  | Shift(_,x,_)  -> collect_expr env x
+  | Shuffle(v,_)  -> collect_var env v
+  | Bitmask(e,_)  -> collect_expr env e
+  | Pack(e1,e2,_) -> collect_expr env e1; collect_expr env e2
+  | Fun(_,l)      -> List.iter (collect_expr env) l
+  | Fun_v(_,_,l)  -> List.iter (collect_expr env) l
 
 let clean_in_deqs (vars:p) (deqs:deq list) : p =
   let env = Hashtbl.create 100 in
diff --git a/src/normalization/expand_array.ml b/src/normalization/expand_array.ml
@@ -144,8 +144,8 @@ let rec expand_expr env_var env_keep env env_it force (e:expr) : expr =
      Shift(op,e1',simpl_arith env ae)
   | Shuffle(v,pat) -> Tuple(List.map (fun x -> Shuffle(x,pat))
                                      (expand_var env_var env_keep env force v))
-  | Mask(e,i) -> Mask(rec_call e,i)
-  | Pack(l,t) -> Pack(List.map rec_call l,t)
+  | Bitmask(e,ae) -> Bitmask(rec_call e,ae)
+  | Pack(e1,e2,t) -> Pack(rec_call e1, rec_call e2, t)
   | Fun(f,el) ->
      if f.name = "refresh" then
        Fun(f,List.map (expand_expr env_var env_keep env env_it
diff --git a/src/normalization/expand_parameters.ml b/src/normalization/expand_parameters.ml
@@ -117,8 +117,8 @@ module Unroll = struct
     | Arith(op,x,y)  -> Arith(op,unroll_expr env_it x,unroll_expr env_it y)
     | Shift(op,e,ae) -> Shift(op,unroll_expr env_it e,simpl_arith env_it ae)
     | Shuffle(v,l)   -> Shuffle(unroll_var env_it v,l)
-    | Mask(e,i)      -> Mask(unroll_expr env_it e,i)
-    | Pack(l,t)      -> Pack(List.map (unroll_expr env_it) l,t)
+    | Bitmask(e,ae)      -> Bitmask(unroll_expr env_it e,simpl_arith env_it ae)
+    | Pack(e1,e2,t)  -> Pack(unroll_expr env_it e1, unroll_expr env_it e2, t)
     | Fun(f,l)       -> Fun(f,List.map (unroll_expr env_it) l)
     | Fun_v _ -> assert false
 
@@ -205,10 +205,10 @@ let rec propagate_expr (expand_env:(var,var list) Hashtbl.t) (e:expr) : expr =
      Shift(op,propagate_expr expand_env e,ae)
   | Shuffle(v,pat) ->
      Shuffle(List.hd (propagate_var expand_env v),pat)
-  | Mask(e,i) ->
-     Mask(propagate_expr expand_env e, i)
-  | Pack(l,t) ->
-     Pack (List.map (propagate_expr expand_env) l, t)
+  | Bitmask(e,ae) ->
+     Bitmask(propagate_expr expand_env e, ae)
+  | Pack(e1,e2,t) ->
+     Pack (propagate_expr expand_env e1, propagate_expr expand_env e2, t)
   | Fun(x,es) ->
      let l = List.map (propagate_expr expand_env) es in
      (match l with
diff --git a/src/normalization/expand_permut.ml b/src/normalization/expand_permut.ml
@@ -27,8 +27,8 @@ let rec apply_perm_e env_fun env_var (e:expr) : expr =
   | Log(op,x,y)   -> Log(op,apply_perm_e env_fun env_var x,apply_perm_e env_fun env_var y)
   | Arith(op,x,y) -> Arith(op,apply_perm_e env_fun env_var x,apply_perm_e env_fun env_var y)
   | Shift(op,e,n) -> Shift(op,apply_perm_e env_fun env_var e,n)
-  | Mask(e,i)     -> Mask(apply_perm_e env_fun env_var e,i)
-  | Pack(l,t)     -> Pack(List.map (apply_perm_e env_fun env_var) l,t)
+  | Bitmask(e,ae)     -> Bitmask(apply_perm_e env_fun env_var e,ae)
+  | Pack(e1,e2,t) -> Pack(apply_perm_e env_fun env_var e1,apply_perm_e env_fun env_var e2,t)
   | Fun(f,l)      -> let l' = List.map (apply_perm_e env_fun env_var) l in
                      (match env_fetch env_fun f with
                       | Some perm -> Tuple (list_from_perm env_var perm l')
diff --git a/src/normalization/fix_dim.ml b/src/normalization/fix_dim.ml
@@ -117,8 +117,8 @@ let rec dim_expr (v_tgt:var) (dim:int) (size:int) (e:expr) : expr =
   | Arith(op,x,y)  -> Arith(op,dim_expr v_tgt dim size x,dim_expr v_tgt dim size y)
   | Shift(op,e,ae) -> Shift(op,dim_expr v_tgt dim size e,ae)
   | Shuffle(v,l)   -> Shuffle(dim_var v_tgt dim size v,l)
-  | Mask(e,i)      -> Mask(dim_expr v_tgt dim size e,i)
-  | Pack(l,t)      -> Pack(List.map (dim_expr v_tgt dim size) l,t)
+  | Bitmask(e,ae)      -> Bitmask(dim_expr v_tgt dim size e,ae)
+  | Pack(e1,e2,t)  -> Pack(dim_expr v_tgt dim size e1,dim_expr v_tgt dim size e2, t)
   | Fun(f,l)       -> Fun(f,List.map (dim_expr v_tgt dim size) l)
   | Fun_v _ -> assert false
 
diff --git a/src/normalization/mask.ml b/src/normalization/mask.ml
@@ -64,8 +64,8 @@ module Get_consts = struct
     | Arith(_,x,y)  -> List.map2 (fun a b -> a && b) (rec_call x) (rec_call y)
     | Shift(_,e',_) -> rec_call e'
     | Shuffle(v,_)  -> [ var_is_const env_const env_not_const v ]
-    | Mask(e',_)    -> rec_call e'
-    | Pack(l,_)     -> flat_map rec_call l
+    | Bitmask(e',_) -> rec_call e'
+    | Pack(e1,e2,_) -> List.map2 (fun a b -> a && b) (rec_call e1) (rec_call e2)
     | Fun(f,l)      ->
        let params_consts = flat_map rec_call l in
        if f.name = "refresh" then params_consts
@@ -405,8 +405,8 @@ let mask_eqn (env_var:(ident,typ) Hashtbl.t)
      | Log(Xor,ExpVar x,ExpVar y) -> mask_xor env_var env_const orig vl x y
      | Log(And as op,ExpVar x,ExpVar y)
      | Log(Or as op,ExpVar x,ExpVar y)  -> mask_and_or env_var env_const orig vl op x y
-     | Mask(ExpVar v,i) -> (* TODO(Mask) *) assert false
-     | Pack(l,typ)      -> (* TODO(Pack) *) assert false
+     | Bitmask(ExpVar v,i) -> (* TODO(Mask) *) assert false
+     | Pack(e1,e2,typ)  -> (* TODO(Pack) *) assert false
      | _ -> Printf.eprintf "Cannot mask expression: %s.\n"
                            (Usuba_print.expr_to_str e);
             assert false
diff --git a/src/normalization/monomorphize.ml b/src/normalization/monomorphize.ml
@@ -225,8 +225,9 @@ module Bslice = struct
     | Shift(op,ExpVar x, ae) ->
        Shift(op,ExpVar(specialize_var env_var x),ae)
     | Shuffle(v,l) -> Shuffle(specialize_var env_var v,l)
-    | Mask(ExpVar v,i) -> Mask(ExpVar (specialize_var env_var v), i)
-    | Pack(l,typ) -> Pack(List.map (specialize_expr env_dir env_m env_var) l,typ)
+    | Bitmask(ExpVar v,ae) -> Bitmask(ExpVar (specialize_var env_var v), ae)
+    | Pack(e1,e2,typ) -> Pack(specialize_expr env_dir env_m env_var e1,
+                              specialize_expr env_dir env_m env_var e2, typ)
     | _ -> Printf.eprintf "Monomorphize::Bslice::specialize_expr: Invalid expr: %s\n"
              (Usuba_print.expr_to_str e);
            assert false
@@ -396,7 +397,7 @@ let rec node_call_can_be_optimized
     | Arith _    -> false
     | Shift _    -> false
     | Shuffle _  -> false
-    | Mask _     -> false
+    | Bitmask _  -> false
     | Pack _     -> false
     | Fun(f',l) ->
        if f'.name = "refresh" then true
diff --git a/src/normalization/norm_tuples.ml b/src/normalization/norm_tuples.ml
@@ -39,8 +39,9 @@ module Simplify_tuples = struct
     | Arith(op,x,y) -> [ Arith(op, expr_from_list (simpl_tuple x),
                                expr_from_list (simpl_tuple y)) ]
     | Shift(op,e,n) -> [ Shift(op,expr_from_list (simpl_tuple e),n) ]
-    | Mask(e,i)     -> [ Mask(expr_from_list (simpl_tuple e),i) ]
-    | Pack(l,t)     -> [ Pack(flat_map simpl_tuple l,t) ]
+    | Bitmask(e,ae)     -> [ Bitmask(expr_from_list (simpl_tuple e),ae) ]
+    | Pack(e1,e2,t) -> [ Pack(expr_from_list (simpl_tuple e1),
+                              expr_from_list (simpl_tuple e2), t) ]
     | Fun(f,l)      ->
        (* If |l| is a Tuple, then the reccursive call goes into Tuple,
           effectively removing the Tuple. *)
diff --git a/src/normalization/remove_sync.ml b/src/normalization/remove_sync.ml
@@ -125,8 +125,8 @@ let rec clean_expr (env_var : (ident,typ) Hashtbl.t)
   | Shuffle(v,l)    -> (match clean_var env_var env_replace v with
                         | [ x ] -> Shuffle(x, l)
                         | xs -> Tuple(List.map (fun x -> Shuffle(x,l)) xs))
-  | Mask(e,i)       -> Mask(rec_call e,i)
-  | Pack(l,t)       -> Pack(List.map rec_call l,t)
+  | Bitmask(e,ae)   -> Bitmask(rec_call e,ae)
+  | Pack(e1,e2,t)   -> Pack(rec_call e1, rec_call e2, t)
   | Fun(f,l)        -> Fun(f,List.map rec_call l)
   | _               -> assert false
 
diff --git a/src/normalization/shift_tuples.ml b/src/normalization/shift_tuples.ml
@@ -72,8 +72,8 @@ let rec shift_expr (env_var:(ident,typ) Hashtbl.t) (e:expr) : expr =
                shift will be performed. *)
             Shift(op,e',n))
             | _ -> Shift(op,e',n))
-  | Mask(e,i) -> Mask(shift_expr env_var e,i)
-  | Pack(l,t) -> Pack(List.map (shift_expr env_var) l,t)
+  | Bitmask(e,ae) -> Bitmask(shift_expr env_var e,ae) (* TODO: something to do here? *)
+  | Pack(e1,e2,t) -> Pack(shift_expr env_var e1, shift_expr env_var e2, t)
   | Fun(f,l)  -> Fun(f,List.map (shift_expr env_var) l)
   | Fun_v(f,ei,l) -> Fun_v(f,ei,List.map (shift_expr env_var) l)
   | _ -> e
diff --git a/src/normalization/unfold_unnest.ml b/src/normalization/unfold_unnest.ml
@@ -49,8 +49,8 @@ let rec get_expr_type env_fun env_var (ltyp:typ list) (e:expr) : typ list =
   | Arith(_,e,_)      -> get_expr_type env_fun env_var ltyp e
   | Shift(_,e,_)      -> get_expr_type env_fun env_var ltyp e
   | Shuffle(v,_)      -> [ get_var_type env_var v ]
-  | Mask(e,_)         -> get_expr_type env_fun env_var ltyp e
-  | Pack(_,Some typ)  -> [ typ ]
+  | Bitmask(e,_)         -> get_expr_type env_fun env_var ltyp e
+  | Pack(_,_,Some typ)  -> [ typ ]
   | Fun(f,l) ->
      if f.name = "rand" then [ Uint(default_dir,Mint 1,1) ]
      else if f.name = "refresh" then
@@ -137,8 +137,8 @@ let rec expand_expr env_var (e:expr) : expr list =
       with Not_found -> (* |ae| does not simplify to a Const *)
         [ Shift(op,x',ae) ])
   | Shuffle(v,pat) -> List.map (fun x -> Shuffle(x,pat)) (expand_var env_var v)
-  | Mask _ -> assert false (* TODO(Mask) *)
-  | Pack _ -> assert false (* TODO(Pack) *)
+  | Bitmask _ -> [ e ] (* TODO: is that correct? *)
+  | Pack _ -> [ e ] (* TODO: is that correct? *)
   | Fun(f,l) -> if f.name = "refresh" && List.length l <> 1 then
                   flat_map (fun v -> expand_expr env_var (Fun(f,[v])))
                            (flat_map (expand_expr env_var) l)
@@ -277,8 +277,26 @@ and norm_expr env_var env_fun (its:(ident*int) list)
   | Shift(op,e,n) ->
      let (deqs,e') = remove_call env_var env_fun its (dir,mtyp) ltyp e in
      deqs, Shift(op,e',n)
-  | Mask _ -> assert false (* TODO(Mask) *)
-  | Pack _ -> assert false (* TODO(Pack) *)
+  | Bitmask(e,i) ->
+     let (deqs,e') = remove_call env_var env_fun its (dir,mtyp) ltyp e in
+     deqs,
+     ( match e' with
+       | Tuple [ e' ] -> Bitmask(e', i)
+       | Tuple _ -> assert false
+       | _ -> Bitmask(e', i) )
+  | Pack(e1,e2,typ) ->
+     let (deqs1, e1') = remove_call env_var env_fun its (dir,mtyp) ltyp e1 in
+     let (deqs2, e2') = remove_call env_var env_fun its (dir,mtyp) ltyp e2 in
+     deqs1 @ deqs2,
+     let e1' = match e1' with
+       | Tuple [ e' ] -> e'
+       | Tuple _ -> assert false
+       | _ -> e1' in
+     let e2' = match e2' with
+       | Tuple [ e' ] -> e'
+       | Tuple _ -> assert false
+       | _ -> e2' in
+     Pack(e1',e2',typ)
   | Fun(f,l) ->
      if f.name = "refresh" && List.length l > 1 then
        norm_expr env_var env_fun its (dir,mtyp) ltyp
diff --git a/src/normalization/unroll.ml b/src/normalization/unroll.ml
@@ -39,8 +39,8 @@ let rec unroll_in_expr (env_it:(ident,int) Hashtbl.t) (e:expr) : expr =
   | Arith(op,x,y)   -> Arith(op,unroll_in_expr env_it x,unroll_in_expr env_it y)
   | Shift(op,e',ae) -> Shift(op,unroll_in_expr env_it e',simpl_arith env_it ae)
   | Shuffle(v,l)    -> Shuffle(unroll_in_var env_it v,l)
-  | Mask(e,i)       -> Mask(unroll_in_expr env_it e,i)
-  | Pack(l,t)       -> Pack(List.map (unroll_in_expr env_it) l,t)
+  | Bitmask(e,ae)   -> Bitmask(unroll_in_expr env_it e,simpl_arith env_it ae)
+  | Pack(e1,e2,t)   -> Pack(unroll_in_expr env_it e1, unroll_in_expr env_it e2,t)
   | Fun(f,l)        -> Fun(f,List.map (unroll_in_expr env_it) l)
   | Fun_v(f,x,l)    -> Fun_v(f, simpl_arith env_it x,
                              List.map (unroll_in_expr env_it) l)
diff --git a/src/optimization/CSE.ml b/src/optimization/CSE.ml
@@ -69,6 +69,8 @@ let rec cse_expr (env_expr:(expr,var list) Hashtbl.t) (e:expr) : expr =
      | Shift(op,e',ae) -> Shift(op,cse_expr env_expr e',ae)
      | Log(op,x,y)     -> Log(op,cse_expr env_expr x,cse_expr env_expr y)
      | Arith(op,x,y)   -> Arith(op,cse_expr env_expr x,cse_expr env_expr y)
+     | Bitmask(e',ae)      -> Bitmask(cse_expr env_expr e',ae)
+     | Pack(e1,e2,t)   -> Pack(cse_expr env_expr e1,cse_expr env_expr e2,t)
      | Fun(f,l)        -> Fun(f,List.map (cse_expr env_expr) l)
      | _ -> Printf.eprintf "cse_expr: invalid expr: %s.\n"
               (Usuba_print.expr_to_str e);
diff --git a/src/optimization/constant_folding.ml b/src/optimization/constant_folding.ml
@@ -172,7 +172,9 @@ let rec fold_expr (env_var:(ident,typ) Hashtbl.t) (e:expr) : expr =
   | Log(op,e1,e2)   -> fold_log env_var op (fold_expr env_var e1)
                                 (fold_expr env_var e2)
   | Arith(op,e1,e2) -> fold_arith env_var op (fold_expr env_var e1)
-                                  (fold_expr env_var e2)
+                         (fold_expr env_var e2)
+  | Bitmask(e',ae)      -> Bitmask(fold_expr env_var e', ae)
+  | Pack(e1,e2,t)   -> Pack(fold_expr env_var e1,fold_expr env_var e2,t)
   | Fun(f,l)        -> Fun(f,List.map (fold_expr env_var) l)
   | _ -> Printf.eprintf "fold_expr: Cannot fold unnormalized expression: %s.\n"
                         (Usuba_print.expr_to_str e);
diff --git a/src/optimization/copy_propagation.ml b/src/optimization/copy_propagation.ml
@@ -139,6 +139,9 @@ module Compute_keeps = struct
                         compute_keep_expr env_keep env_var y
     | Arith(_,x,y)   -> compute_keep_expr env_keep env_var x;
                         compute_keep_expr env_keep env_var y
+    | Bitmask(e',_)     -> compute_keep_expr env_keep env_var e'
+    | Pack(e1,e2,_)  -> compute_keep_expr env_keep env_var e1;
+                        compute_keep_expr env_keep env_var e2
     | _ -> Printf.eprintf "compute_keep_expr: invalid expr: %s.\n"
              (Usuba_print.expr_to_str e);
            assert false
@@ -265,7 +268,11 @@ module Backwards_cp = struct
     | Log(op,x,y)     -> Log(op,propagate_in_expr optimized_away x,
                              propagate_in_expr optimized_away y)
     | Arith(op,x,y)   -> Arith(op,propagate_in_expr optimized_away x,
-                                   propagate_in_expr optimized_away y)
+                               propagate_in_expr optimized_away y)
+    | Bitmask(e',ae)      -> Bitmask(propagate_in_expr optimized_away e',
+                               propagate_in_aexpr optimized_away ae)
+    | Pack(e1,e2,t)   -> Pack(propagate_in_expr optimized_away e1,
+                              propagate_in_expr optimized_away e2, t)
     | Fun(f,l)        -> Fun(f,(List.map (propagate_in_expr optimized_away) l))
     | _ -> Printf.eprintf "propagate_in_expr: invalid expr: %s.\n"
                           (Usuba_print.expr_to_str e);
@@ -386,7 +393,11 @@ let rec propagate_in_expr  (optimized_away:(var,expr*((ident*deq_i) list)) Hasht
   | Log(op,x,y)     -> [], Log(op,propagate_in_expr_rec optimized_away x,
                                  propagate_in_expr_rec optimized_away y)
   | Arith(op,x,y)   -> [], Arith(op,propagate_in_expr_rec optimized_away x,
-                                   propagate_in_expr_rec optimized_away y)
+                                 propagate_in_expr_rec optimized_away y)
+  | Bitmask(e',ae)      -> [], Bitmask(propagate_in_expr_rec optimized_away e',
+                                       propagate_in_aexpr optimized_away ae)
+  | Pack(e1,e2,t)   -> [], Pack(propagate_in_expr_rec optimized_away e1,
+                                propagate_in_expr_rec optimized_away e2, t)
   | Fun(f,l)        -> [], Fun(f,(List.map (propagate_in_expr_rec optimized_away) l))
   | _ -> Printf.eprintf "propagate_in_expr: invalid expr: %s.\n"
            (Usuba_print.expr_to_str e);
diff --git a/src/optimization/get_live_var.ml b/src/optimization/get_live_var.ml
@@ -16,6 +16,8 @@ let rec add_right env_var live e =
   | Shuffle(v,_) -> List.iter (fun v -> Hashtbl.replace live (get_var_base v) true)
                               (expand_var env_var v)
   | Arith(_,x,y) -> add_right env_var live x; add_right env_var live y
+  | Bitmask(e',_) -> add_right env_var live e'
+  | Pack(e1,e2,_) -> add_right env_var live e1; add_right env_var live e2
   | Fun(_,l) -> List.iter (add_right env_var live) l
   | _ -> assert false
 
diff --git a/src/optimization/inline.ml b/src/optimization/inline.ml
@@ -40,8 +40,8 @@ module Is_linear = struct
                         | Mul | Div | Mod -> false)
     | Shift(_,e',_) -> is_linear_expr env_fun e'
     | Shuffle _     -> true
-    | Mask(e',_)    -> is_linear_expr env_fun e'
-    | Pack(l,_)     -> List.for_all (is_linear_expr env_fun) l
+    | Bitmask(e',_)    -> is_linear_expr env_fun e'
+    | Pack(e1,e2,_) -> (is_linear_expr env_fun e1) && (is_linear_expr env_fun e2)
     | Fun(f,l)      -> (is_linear_def_by_name env_fun f)
                        && (List.for_all (is_linear_expr env_fun) l)
     | Fun_v(f,_,l)  -> (is_linear_def_by_name env_fun f)
diff --git a/src/optimization/linearize_arrays.ml b/src/optimization/linearize_arrays.ml
@@ -124,6 +124,9 @@ let rec replace_expr to_linearize (e:expr) : expr =
   | Arith(op,x,y)  -> Arith(op,replace_expr to_linearize x,
                             replace_expr to_linearize y)
   | Shuffle(v,pat) -> Shuffle(replace_var to_linearize v,pat)
+  | Bitmask(e',ae) -> Bitmask(replace_expr to_linearize e', ae)
+  | Pack(e1,e2,t)  -> Pack(replace_expr to_linearize e1,
+                           replace_expr to_linearize e2, t)
   | Fun(f,l)       -> Fun(f,List.map (replace_expr to_linearize) l)
   | _ -> assert false
 
diff --git a/src/optimization/vital_inline.ml b/src/optimization/vital_inline.ml
diff --git a/src/parsing/lexer.mll b/src/parsing/lexer.mll
diff --git a/src/parsing/parser.mly b/src/parsing/parser.mly
diff --git a/src/rename.ml b/src/rename.ml
diff --git a/src/usuba_AST.v b/src/usuba_AST.v
diff --git a/src/usuba_print.ml b/src/usuba_print.ml
diff --git a/src/utils.ml b/src/utils.ml
diff --git a/src/verification/type_checker.ml b/src/verification/type_checker.ml
