diff --git a/src/compiler/nir/nir_lower_alpha_test.c b/src/compiler/nir/nir_lower_alpha_test.c
index 78a1abd9b8a..86105c87391 100644
--- a/src/compiler/nir/nir_lower_alpha_test.c
+++ b/src/compiler/nir/nir_lower_alpha_test.c
@@ -44,8 +44,7 @@ nir_lower_alpha_test(nir_shader *shader, enum compare_func func,
    assert(shader->info.stage == MESA_SHADER_FRAGMENT);
 
    nir_foreach_function_impl(impl, shader) {
-      nir_builder b = nir_builder_create(impl);
-      b.cursor = nir_before_cf_list(&impl->body);
+      nir_builder b = nir_builder_at(nir_before_cf_list(&impl->body));
 
       nir_foreach_block(block, impl) {
          nir_foreach_instr_safe(instr, block) {
diff --git a/src/compiler/nir/nir_lower_bitmap.c b/src/compiler/nir/nir_lower_bitmap.c
index d717410e032..95999860de8 100644
--- a/src/compiler/nir/nir_lower_bitmap.c
+++ b/src/compiler/nir/nir_lower_bitmap.c
@@ -103,8 +103,7 @@ static void
 lower_bitmap_impl(nir_function_impl *impl,
                   const nir_lower_bitmap_options *options)
 {
-   nir_builder b = nir_builder_create(impl);
-   b.cursor = nir_before_cf_list(&impl->body);
+   nir_builder b = nir_builder_at(nir_before_cf_list(&impl->body));
 
    lower_bitmap(impl->function->shader, &b, options);
 
diff --git a/src/compiler/nir/nir_lower_clip.c b/src/compiler/nir/nir_lower_clip.c
index c87da9dc678..80fc5ad1950 100644
--- a/src/compiler/nir/nir_lower_clip.c
+++ b/src/compiler/nir/nir_lower_clip.c
@@ -428,8 +428,7 @@ lower_clip_fs(nir_function_impl *impl, unsigned ucp_enables,
               nir_variable **in, bool use_clipdist_array)
 {
    nir_ssa_def *clipdist[MAX_CLIP_PLANES];
-   nir_builder b = nir_builder_create(impl);
-   b.cursor = nir_before_cf_list(&impl->body);
+   nir_builder b = nir_builder_at(nir_before_cf_list(&impl->body));
 
    if (!use_clipdist_array) {
       if (ucp_enables & 0x0f)
diff --git a/src/compiler/nir/nir_lower_goto_ifs.c b/src/compiler/nir/nir_lower_goto_ifs.c
index 824f11d5b0e..50f42971d15 100644
--- a/src/compiler/nir/nir_lower_goto_ifs.c
+++ b/src/compiler/nir/nir_lower_goto_ifs.c
@@ -944,8 +944,7 @@ nir_lower_goto_ifs_impl(nir_function_impl *impl)
    /* From this point on, it's structured */
    impl->structured = true;
 
-   nir_builder b = nir_builder_create(impl);
-   b.cursor = nir_before_block(nir_start_block(impl));
+   nir_builder b = nir_builder_at(nir_before_block(nir_start_block(impl)));
 
    void *mem_ctx = ralloc_context(b.shader);
 
diff --git a/src/compiler/nir/nir_lower_gs_intrinsics.c b/src/compiler/nir/nir_lower_gs_intrinsics.c
index 5f07c907e6a..e8f0fdf394d 100644
--- a/src/compiler/nir/nir_lower_gs_intrinsics.c
+++ b/src/compiler/nir/nir_lower_gs_intrinsics.c
@@ -394,11 +394,9 @@ nir_lower_gs_intrinsics(nir_shader *shader, nir_lower_gs_intrinsics_flags option
    if (!a_block_needs_set_vertex_and_primitive_count(impl->end_block, per_stream))
       return false;
 
-   nir_builder b = nir_builder_create(impl);
+   nir_builder b = nir_builder_at(nir_before_cf_list(&impl->body));
    state.builder = &b;
 
-   b.cursor = nir_before_cf_list(&impl->body);
-
    for (unsigned i = 0; i < NIR_MAX_XFB_STREAMS; i++) {
       if (per_stream && !(shader->info.gs.active_stream_mask & (1 << i)))
          continue;
diff --git a/src/compiler/nir/nir_lower_io_to_temporaries.c b/src/compiler/nir/nir_lower_io_to_temporaries.c
index 08af9670920..6d889de41e4 100644
--- a/src/compiler/nir/nir_lower_io_to_temporaries.c
+++ b/src/compiler/nir/nir_lower_io_to_temporaries.c
@@ -276,8 +276,7 @@ static void
 emit_input_copies_impl(struct lower_io_state *state, nir_function_impl *impl)
 {
    if (impl == state->entrypoint) {
-      nir_builder b = nir_builder_create(impl);
-      b.cursor = nir_before_block(nir_start_block(impl));
+      nir_builder b = nir_builder_at(nir_before_block(nir_start_block(impl)));
       emit_copies(&b, &state->old_inputs, &state->new_inputs);
       if (state->shader->info.stage == MESA_SHADER_FRAGMENT)
          fixup_interpolation(state, impl, &b);
diff --git a/src/compiler/nir/nir_lower_is_helper_invocation.c b/src/compiler/nir/nir_lower_is_helper_invocation.c
index dcddb0777ee..bce5999a139 100644
--- a/src/compiler/nir/nir_lower_is_helper_invocation.c
+++ b/src/compiler/nir/nir_lower_is_helper_invocation.c
@@ -103,8 +103,7 @@ nir_lower_is_helper_invocation(nir_shader *shader)
 
    nir_function_impl *entrypoint = nir_shader_get_entrypoint(shader);
 
-   nir_builder b = nir_builder_create(entrypoint);
-   b.cursor = nir_before_cf_list(&entrypoint->body);
+   nir_builder b = nir_builder_at(nir_before_cf_list(&entrypoint->body));
 
    nir_variable *is_helper = nir_local_variable_create(entrypoint,
                                           glsl_bool_type(),
diff --git a/src/compiler/nir/nir_lower_load_const_to_scalar.c b/src/compiler/nir/nir_lower_load_const_to_scalar.c
index 415212362d1..0cd7a7d3372 100644
--- a/src/compiler/nir/nir_lower_load_const_to_scalar.c
+++ b/src/compiler/nir/nir_lower_load_const_to_scalar.c
@@ -41,8 +41,7 @@ lower_load_const_instr_scalar(nir_load_const_instr *lower)
    if (lower->def.num_components == 1)
       return false;
 
-   nir_builder b = nir_builder_create(nir_cf_node_get_function(&lower->instr.block->cf_node));
-   b.cursor = nir_before_instr(&lower->instr);
+   nir_builder b = nir_builder_at(nir_before_instr(&lower->instr));
 
    /* Emit the individual loads. */
    nir_ssa_def *loads[NIR_MAX_VEC_COMPONENTS];
diff --git a/src/compiler/nir/nir_lower_multiview.c b/src/compiler/nir/nir_lower_multiview.c
index 4ff6fc23ff4..8201c3f2839 100644
--- a/src/compiler/nir/nir_lower_multiview.c
+++ b/src/compiler/nir/nir_lower_multiview.c
@@ -203,8 +203,7 @@ nir_lower_multiview(nir_shader *shader, uint32_t view_mask)
    nir_cf_list body;
    nir_cf_list_extract(&body, &entrypoint->body);
 
-   nir_builder b = nir_builder_create(entrypoint);
-   b.cursor = nir_after_cf_list(&entrypoint->body);
+   nir_builder b = nir_builder_at(nir_after_cf_list(&entrypoint->body));
 
    /* Loop Index will go from 0 to view_count. */
    nir_variable *loop_index_var =
diff --git a/src/compiler/nir/nir_lower_passthrough_edgeflags.c b/src/compiler/nir/nir_lower_passthrough_edgeflags.c
index 2c47deb7509..32db90f34b0 100644
--- a/src/compiler/nir/nir_lower_passthrough_edgeflags.c
+++ b/src/compiler/nir/nir_lower_passthrough_edgeflags.c
@@ -32,8 +32,7 @@ lower_impl(nir_function_impl *impl)
    nir_variable *in, *out;
    nir_ssa_def *def;
 
-   b = nir_builder_create(impl);
-   b.cursor = nir_before_cf_list(&impl->body);
+   b = nir_builder_at(nir_before_cf_list(&impl->body));
 
    /* The edge flag is the last input in st/mesa.  This code is also called by
     * i965 which calls it before any input locations are assigned.
diff --git a/src/compiler/nir/nir_lower_shader_calls.c b/src/compiler/nir/nir_lower_shader_calls.c
index 06dcf648c1c..78eb69b12ea 100644
--- a/src/compiler/nir/nir_lower_shader_calls.c
+++ b/src/compiler/nir/nir_lower_shader_calls.c
@@ -1102,8 +1102,7 @@ flatten_resume_if_ladder(nir_builder *b,
             /* We want to place anything re-materialized from inside the loop
              * at the top of the resume half of the loop.
              */
-            nir_builder bl = nir_builder_create(b->impl);
-            bl.cursor = nir_before_cf_list(&_if->then_list);
+            nir_builder bl = nir_builder_at(nir_before_cf_list(&_if->then_list));
 
             ASSERTED bool found =
                flatten_resume_if_ladder(&bl, &_if->cf_node, &_if->then_list,
@@ -1273,8 +1272,7 @@ lower_resume(nir_shader *shader, int call_idx)
    /* Create a nop instruction to use as a cursor as we extract and re-insert
     * stuff into the CFG.
     */
-   nir_builder b = nir_builder_create(impl);
-   b.cursor = nir_before_cf_list(&impl->body);
+   nir_builder b = nir_builder_at(nir_before_cf_list(&impl->body));
    ASSERTED bool found =
       flatten_resume_if_ladder(&b, &impl->cf_node, &impl->body,
                                true, resume_instr, &remat);
@@ -1517,8 +1515,7 @@ nir_opt_trim_stack_values(nir_shader *shader)
             continue;
          }
 
-         nir_builder b = nir_builder_create(impl);
-         b.cursor = nir_before_instr(instr);
+         nir_builder b = nir_builder_at(nir_before_instr(instr));
 
          nir_ssa_def *value = nir_channels(&b, intrin->src[0].ssa, read_mask);
          nir_instr_rewrite_src_ssa(instr, &intrin->src[0], value);
diff --git a/src/compiler/nir/nir_lower_texcoord_replace.c b/src/compiler/nir/nir_lower_texcoord_replace.c
index 642eff95444..9b342b57468 100644
--- a/src/compiler/nir/nir_lower_texcoord_replace.c
+++ b/src/compiler/nir/nir_lower_texcoord_replace.c
@@ -63,8 +63,7 @@ nir_lower_texcoord_replace_impl(nir_function_impl *impl,
                                 bool point_coord_is_sysval,
                                 bool yinvert)
 {
-   nir_builder b = nir_builder_create(impl);
-   b.cursor = nir_before_cf_list(&impl->body);
+   nir_builder b = nir_builder_at(nir_before_cf_list(&impl->body));
 
    nir_ssa_def *new_coord;
    if (point_coord_is_sysval) {
diff --git a/src/compiler/nir/nir_lower_variable_initializers.c b/src/compiler/nir/nir_lower_variable_initializers.c
index b1f9b1f045a..bd128eeb9f7 100644
--- a/src/compiler/nir/nir_lower_variable_initializers.c
+++ b/src/compiler/nir/nir_lower_variable_initializers.c
@@ -145,8 +145,8 @@ nir_zero_initialize_shared_memory(nir_shader *shader,
    assert(chunk_size > 0);
    assert(chunk_size % 4 == 0);
 
-   nir_builder b = nir_builder_create(nir_shader_get_entrypoint(shader));
-   b.cursor = nir_before_cf_list(&b.impl->body);
+   nir_function_impl *impl = nir_shader_get_entrypoint(shader);
+   nir_builder b = nir_builder_at(nir_before_cf_list(&impl->body));
 
    assert(!shader->info.workgroup_size_variable);
    const unsigned local_count = shader->info.workgroup_size[0] *
diff --git a/src/compiler/nir/nir_opt_copy_propagate.c b/src/compiler/nir/nir_opt_copy_propagate.c
index 69da45156cf..753bd7090e6 100644
--- a/src/compiler/nir/nir_opt_copy_propagate.c
+++ b/src/compiler/nir/nir_opt_copy_propagate.c
@@ -57,13 +57,12 @@ is_swizzleless_move(nir_alu_instr *instr)
 }
 
 static bool
-rewrite_to_vec(nir_function_impl *impl, nir_alu_instr *mov, nir_alu_instr *vec)
+rewrite_to_vec(nir_alu_instr *mov, nir_alu_instr *vec)
 {
    if (mov->op != nir_op_mov)
       return false;
 
-   nir_builder b = nir_builder_create(impl);
-   b.cursor = nir_after_instr(&mov->instr);
+   nir_builder b = nir_builder_at(nir_after_instr(&mov->instr));
 
    unsigned num_comp = mov->dest.dest.ssa.num_components;
    nir_alu_instr *new_vec = nir_alu_instr_create(b.shader, nir_op_vec(num_comp));
@@ -80,7 +79,7 @@ rewrite_to_vec(nir_function_impl *impl, nir_alu_instr *mov, nir_alu_instr *vec)
 }
 
 static bool
-copy_propagate_alu(nir_function_impl *impl, nir_alu_src *src, nir_alu_instr *copy)
+copy_propagate_alu(nir_alu_src *src, nir_alu_instr *copy)
 {
    nir_ssa_def *def = NULL;
    nir_alu_instr *user = nir_instr_as_alu(src->src.parent_instr);
@@ -98,7 +97,7 @@ copy_propagate_alu(nir_function_impl *impl, nir_alu_src *src, nir_alu_instr *cop
 
       for (unsigned i = 1; i < num_comp; i++) {
          if (copy->src[src->swizzle[i]].src.ssa != def)
-            return rewrite_to_vec(impl, user, copy);
+            return rewrite_to_vec(user, copy);
       }
 
       for (unsigned i = 0; i < num_comp; i++)
@@ -122,7 +121,7 @@ copy_propagate(nir_src *src, nir_alu_instr *copy)
 }
 
 static bool
-copy_prop_instr(nir_function_impl *impl, nir_instr *instr)
+copy_prop_instr(nir_instr *instr)
 {
    if (instr->type != nir_instr_type_alu)
       return false;
@@ -136,7 +135,7 @@ copy_prop_instr(nir_function_impl *impl, nir_instr *instr)
 
    nir_foreach_use_including_if_safe(src, &mov->dest.dest.ssa) {
       if (!src->is_if && src->parent_instr->type == nir_instr_type_alu)
-         progress |= copy_propagate_alu(impl, container_of(src, nir_alu_src, src), mov);
+         progress |= copy_propagate_alu(container_of(src, nir_alu_src, src), mov);
       else
          progress |= copy_propagate(src, mov);
    }
@@ -154,7 +153,7 @@ nir_copy_prop_impl(nir_function_impl *impl)
 
    nir_foreach_block(block, impl) {
       nir_foreach_instr_safe(instr, block) {
-         progress |= copy_prop_instr(impl, instr);
+         progress |= copy_prop_instr(instr);
       }
    }
 
diff --git a/src/compiler/nir/nir_opt_load_store_vectorize.c b/src/compiler/nir/nir_opt_load_store_vectorize.c
index c8cdc7819ec..90abbcc179f 100644
--- a/src/compiler/nir/nir_opt_load_store_vectorize.c
+++ b/src/compiler/nir/nir_opt_load_store_vectorize.c
@@ -1140,7 +1140,7 @@ try_vectorize(nir_function_impl *impl, struct vectorize_ctx *ctx,
 }
 
 static bool
-try_vectorize_shared2(nir_function_impl *impl, struct vectorize_ctx *ctx,
+try_vectorize_shared2(struct vectorize_ctx *ctx,
                       struct entry *low, struct entry *high,
                       struct entry *first, struct entry *second)
 {
@@ -1178,9 +1178,7 @@ try_vectorize_shared2(nir_function_impl *impl, struct vectorize_ctx *ctx,
    }
 
    /* vectorize the accesses */
-   nir_builder b = nir_builder_create(impl);
-
-   b.cursor = nir_after_instr(first->is_store ? second->instr : first->instr);
+   nir_builder b = nir_builder_at(nir_after_instr(first->is_store ? second->instr : first->instr));
 
    nir_ssa_def *offset = first->intrin->src[first->is_store].ssa;
    offset = nir_iadd_imm(&b, offset, nir_intrinsic_base(first->intrin));
@@ -1247,7 +1245,7 @@ vectorize_sorted_entries(struct vectorize_ctx *ctx, nir_function_impl *impl,
                 get_variable_mode(first) != nir_var_mem_shared)
                break;
 
-            if (try_vectorize_shared2(impl, ctx, low, high, first, second)) {
+            if (try_vectorize_shared2(ctx, low, high, first, second)) {
                low = NULL;
                *util_dynarray_element(arr, struct entry *, second_idx) = NULL;
                progress = true;
diff --git a/src/compiler/nir/nir_opt_peephole_select.c b/src/compiler/nir/nir_opt_peephole_select.c
index 091416824ea..288ec6aaa4f 100644
--- a/src/compiler/nir/nir_opt_peephole_select.c
+++ b/src/compiler/nir/nir_opt_peephole_select.c
@@ -356,8 +356,7 @@ nir_opt_collapse_if(nir_if *if_stmt, nir_shader *shader, unsigned limit,
    }
 
    /* combine the conditions */
-   struct nir_builder b = nir_builder_create(nir_cf_node_get_function(&if_stmt->cf_node)->function->impl);
-   b.cursor = nir_before_cf_node(&if_stmt->cf_node);
+   struct nir_builder b = nir_builder_at(nir_before_cf_node(&if_stmt->cf_node));
    nir_ssa_def *cond = nir_iand(&b, if_stmt->condition.ssa,
                                 parent_if->condition.ssa);
    nir_if_rewrite_condition(if_stmt, nir_src_for_ssa(cond));
diff --git a/src/compiler/nir/nir_opt_vectorize.c b/src/compiler/nir/nir_opt_vectorize.c
index 832a738ca0d..56dce8e4e7d 100644
--- a/src/compiler/nir/nir_opt_vectorize.c
+++ b/src/compiler/nir/nir_opt_vectorize.c
@@ -201,8 +201,7 @@ instr_try_combine(struct set *instr_set, nir_instr *instr1, nir_instr *instr2)
    if (total_components > instr1->pass_flags)
       return NULL;
 
-   nir_builder b = nir_builder_create(nir_cf_node_get_function(&instr1->block->cf_node));
-   b.cursor = nir_after_instr(instr1);
+   nir_builder b = nir_builder_at(nir_after_instr(instr1));
 
    nir_alu_instr *new_alu = nir_alu_instr_create(b.shader, alu1->op);
    nir_ssa_dest_init(&new_alu->instr, &new_alu->dest.dest, total_components,