From 142728671691ee3cff26b33afd789e130ad1c183 Mon Sep 17 00:00:00 2001 From: Tarry Singh Date: Tue, 15 Jul 2025 15:26:03 +0000 Subject: [PATCH] runtimes/neuron: fix neuron runtime This PR fixes the neuron runtime with the following: Proxy the PJRT Api method to enforce the client struct sizes since the neuron PJRT plugin doesn't use `>=` but `==` to assert them, breaking PJRT compatibility guarantees. Fixes https://github.com/aws-neuron/aws-neuron-sdk/issues/1095 Reimplement `libneuronxla` in Zig to control neuronx-cc sandboxing and invocation. Implement a python bootstrapper in Zig to create a full blown `neuronx-cc` executable, avoiding the infamous chicken and egg problem of python executables boostrapping when sandboxed (due to fixed path shebangs). --------- Co-authored-by: Corentin Kerisit --- MODULE.bazel | 3 + runtimes/neuron/BUILD.bazel | 161 +++++------ runtimes/neuron/empty.py | 0 runtimes/neuron/libneuronxla.zig | 249 +++++++++++++++++ runtimes/neuron/libpjrt_neuron.BUILD.bazel | 70 ++--- runtimes/neuron/libpjrt_neuron.zig | 227 ++++++++++++--- runtimes/neuron/neuron.bzl | 2 +- runtimes/neuron/neuron.zig | 99 +------ runtimes/neuron/neuronx-cc.zig | 89 ++++++ runtimes/neuron/{libpjrt_neuron.h => nrt.h} | 0 runtimes/neuron/pyenv.bzl | 41 --- runtimes/neuron/requirements.lock.txt | 294 ++++++++++---------- runtimes/neuron/zmlxneuron.c | 24 -- upb/BUILD.bazel | 8 +- upb/upb.h | 1 + upb/upb.zig | 10 +- 16 files changed, 813 insertions(+), 465 deletions(-) delete mode 100644 runtimes/neuron/empty.py create mode 100644 runtimes/neuron/libneuronxla.zig create mode 100644 runtimes/neuron/neuronx-cc.zig rename runtimes/neuron/{libpjrt_neuron.h => nrt.h} (100%) delete mode 100644 runtimes/neuron/pyenv.bzl delete mode 100644 runtimes/neuron/zmlxneuron.c create mode 100644 upb/upb.h diff --git a/MODULE.bazel b/MODULE.bazel index 105eeca..0923038 100644 --- a/MODULE.bazel +++ b/MODULE.bazel @@ -59,6 +59,9 @@ pip.parse( "--python-version=311", "--platform=linux_x86_64", "--platform=manylinux2014_x86_64", + "--platform=manylinux_2_27_x86_64", + "--platform=manylinux_2_28_x86_64", + "--platform=manylinux_2_29_x86_64", ], hub_name = "neuron_py_deps", python_version = "3.11", diff --git a/runtimes/neuron/BUILD.bazel b/runtimes/neuron/BUILD.bazel index 0fbe667..184925a 100644 --- a/runtimes/neuron/BUILD.bazel +++ b/runtimes/neuron/BUILD.bazel @@ -1,20 +1,8 @@ +load("@com_google_protobuf//bazel:upb_proto_library.bzl", "upb_c_proto_library") load("@rules_cc//cc:cc_library.bzl", "cc_library") -load("@rules_python//python:pip.bzl", "compile_pip_requirements") -load("@rules_python//python/entry_points:py_console_script_binary.bzl", "py_console_script_binary") -load("@rules_uv//uv:pip.bzl", "pip_compile") -load("@rules_zig//zig:defs.bzl", "zig_library", "zig_shared_library") -load(":neuron.bzl", "py_binary_with_script") -load(":pyenv.bzl", "pyenv_zig") - -pyenv_zig( - name = "libneuronxla_pyenv_zig", - deps = [":libneuronxla"], -) - -zig_library( - name = "libneuronxla_pyenv", - main = ":libneuronxla_pyenv_zig", -) +load("@rules_python//python/uv:lock.bzl", uv_lock = "lock") +load("@rules_zig//zig:defs.bzl", "zig_binary", "zig_library", "zig_shared_library") +load("@zml//bazel:runfiles.bzl", "runfiles_to_default") # A proxy PJRT Plugin that loads the Neuron PJRT Plugin # and returns the instance from nested GetPjrtApi. @@ -22,101 +10,107 @@ zig_library( # Additionally, it provides a way to load implicit transitive dependencies # of neuronx-cc (see add_needed of the patchelf target below). zig_shared_library( - name = "libpjrt_neuron_proxy", + name = "libpjrt_neuron", copts = ["-fno-stack-check"], main = "libpjrt_neuron.zig", visibility = ["@libpjrt_neuron//:__subpackages__"], deps = [ + ":libpython", "//stdx", "@rules_zig//zig/runfiles", + "@xla//xla/pjrt/c:pjrt_c_api_hdrs", ], ) -pip_compile( - name = "update_requirements", - args = [ - "--generate-hashes", - "--emit-index-url", - "--emit-find-links", - "--no-strip-extras", - "--index-strategy=unsafe-best-match", - "--upgrade", - ], - python_platform = "x86_64-unknown-linux-gnu", - requirements_in = "requirements.in", - requirements_txt = "requirements.lock.txt", -) - -py_console_script_binary( +zig_binary( name = "neuronx-cc", - binary_rule = py_binary_with_script, - pkg = "@neuron_py_deps//neuronx_cc", - visibility = ["//visibility:public"], + data = ["@neuron_py_deps//neuronx_cc"], + linkopts = ["-Wl,-rpath,$ORIGIN/../lib"], + main = "neuronx-cc.zig", + tags = ["manual"], + deps = [":libpython"], ) -py_binary_with_script( - name = "libneuronxla", - srcs = ["empty.py"], - main = "empty.py", - deps = [ - ":neuronx-cc", - "@neuron_py_deps//libneuronxla", - ], +runfiles_to_default( + name = "neuronx-cc_files", + visibility = ["@libpjrt_neuron//:__subpackages__"], + deps = [":neuronx-cc"], ) cc_library( name = "libpython", hdrs = ["libpython.h"], - visibility = ["//visibility:public"], deps = [ "@rules_python//python/cc:current_py_cc_headers", "@rules_python//python/cc:current_py_cc_libs", ], ) -compile_pip_requirements( - name = "requirements", - src = "requirements.in", - py_binary = py_binary_with_script, - requirements_txt = "requirements.lock.txt", - tags = [ - "manual", - ], -) - -cc_library( - name = "zmlxneuron_lib", - srcs = ["zmlxneuron.c"], - linkopts = ["-ldl"], - visibility = ["@libpjrt_neuron//:__subpackages__"], -) - -alias( - name = "bootstrap", - actual = "@rules_python//python/config_settings:bootstrap_impl", -) - cc_library( name = "empty", ) cc_library( - name = "libpjrt_neuron", - hdrs = ["libpjrt_neuron.h"], + name = "zmlxneuron", defines = ["ZML_RUNTIME_NEURON"], - deps = [ - "@libpjrt_neuron", +) + +cc_library( + name = "libnrt_headers", + hdrs = ["nrt.h"], + deps = ["@libpjrt_neuron//:libnrt_headers"], +) + +filegroup( + name = "layers", + srcs = [], + visibility = ["//visibility:public"], +) + +upb_c_proto_library( + name = "xla_data_upb", + deps = ["@xla//xla:xla_data_proto"], +) + +upb_c_proto_library( + name = "hlo_proto_upb", + deps = ["@xla//xla/service:hlo_proto"], +) + +zig_shared_library( + name = "libneuronxla", + copts = [ + "-fno-stack-check", + "-fPIC", ], + main = "libneuronxla.zig", + shared_lib_name = "libneuronxla.so", + visibility = ["@libpjrt_neuron//:__subpackages__"], + deps = [ + ":hlo_proto_upb", + ":libpython", + ":xla_data_upb", + "//stdx", + "//upb", + ], +) + +uv_lock( + name = "requirements", + srcs = ["requirements.in"], + out = "requirements.lock.txt", + args = [ + "--emit-index-url", + "--emit-find-links", + "--index-strategy=unsafe-best-match", + "--upgrade", + ], + python_version = "3.11", + tags = ["manual"], ) zig_library( name = "neuron", - data = select({ - "//runtimes:neuron.enabled": [ - ":libneuronxla", - ], - "//conditions:default": [], - }), import_name = "runtimes/neuron", main = "neuron.zig", visibility = ["//visibility:public"], @@ -124,21 +118,14 @@ zig_library( "//pjrt", ] + select({ "//runtimes:neuron.enabled": [ - ":libneuronxla_pyenv", - ":libpjrt_neuron", + ":libnrt_headers", ":libpython", + ":zmlxneuron", "//async", "//stdx", + "@libpjrt_neuron", "@rules_zig//zig/runfiles", ], "//conditions:default": [":empty"], }), ) - -filegroup( - name = "layers", - srcs = [ - "@apt_neuron//bash/amd64", - ], - visibility = ["//visibility:public"], -) diff --git a/runtimes/neuron/empty.py b/runtimes/neuron/empty.py deleted file mode 100644 index e69de29..0000000 diff --git a/runtimes/neuron/libneuronxla.zig b/runtimes/neuron/libneuronxla.zig new file mode 100644 index 0000000..ae5710e --- /dev/null +++ b/runtimes/neuron/libneuronxla.zig @@ -0,0 +1,249 @@ +const std = @import("std"); + +const c = @import("c"); +const stdx = @import("stdx"); +const upb = @import("upb"); + +const log = std.log.scoped(.@"zml/runtimes/neuron/libneuronxla"); + +pub fn makeTempDir(buf: []u8, prefix: []const u8) ![]const u8 { + const tmp_dir = std.posix.getenv("TMPDIR") orelse "/tmp"; + const ret = try std.fmt.bufPrint(buf, "{s}{s}{s}{d}", .{ + tmp_dir, + std.fs.path.sep_str_posix, + prefix, + std.time.microTimestamp(), + }); + try std.fs.makeDirAbsolute(ret); + return ret; +} + +var module_def: c.PyModuleDef = .{ + .m_base = .{}, + .m_name = "libneuronxla", + .m_size = 0, + .m_methods = @constCast(&[_]c.PyMethodDef{ + .{ + .ml_name = "hook", + .ml_meth = @ptrCast(&hook), + .ml_flags = c.METH_NOARGS, + .ml_doc = "Return a greeting from Zig.", + }, + .{ + .ml_name = "neuronx_cc", + .ml_meth = @ptrCast(&neuronx_cc), + .ml_flags = c.METH_FASTCALL, + .ml_doc = "Return a greeting from Zig.", + }, + .{}, + }), + .m_slots = @constCast(&[_]c.PyModuleDef_Slot{ + .{ .slot = c.Py_mod_exec, .value = @constCast(@ptrCast(&module_exec)) }, + .{}, + }), + .m_traverse = null, + .m_clear = null, + .m_free = null, +}; + +fn module_exec(module: ?*c.PyObject) callconv(.c) c_int { + _ = module; + return 0; +} + +fn hook(self: ?*c.PyObject, args: ?*c.PyObject) callconv(.c) ?*c.PyObject { + _ = self; + _ = args; + const none = c.Py_None(); + defer c.Py_IncRef(none); + return none; +} + +pub fn PyBytes_AsStringAndSize(object: *c.PyObject) []u8 { + var buf: [*c]u8 = undefined; + var len: c.Py_ssize_t = undefined; + _ = c.PyBytes_AsStringAndSize(object, &buf, &len); + return buf[0..@intCast(len)]; +} + +fn wrapNeffAsCustomCall(allocator: std.mem.Allocator, hlo_code: []const u8, neff_file_path: []const u8) ![]const u8 { + var upb_alloc: upb.Allocator = .init(allocator); + const upb_arena = c.upb_Arena_Init(null, 0, upb_alloc.inner()); + + const hlo_module = try upb.parse(c.xla_HloModuleProto, upb_arena, hlo_code); + + const entry = blk: { + var size: usize = undefined; + const computations = c.xla_HloModuleProto_mutable_computations(hlo_module, &size)[0..size]; + for (computations) |comp| { + if (c.xla_HloComputationProto_id(comp) == c.xla_HloModuleProto_entry_computation_id(hlo_module)) { + break :blk comp; + } + } else return error.ComputationNotFound; + }; + + const entry_instructions = blk: { + var size: usize = undefined; + break :blk c.xla_HloComputationProto_instructions(entry, &size)[0..size]; + }; + c.xla_HloComputationProto_clear_instructions(entry); + + const fused_root = blk: { + for (entry_instructions) |instruction| { + if (c.xla_HloInstructionProto_id(instruction) == c.xla_HloComputationProto_root_id(entry)) { + break :blk try upb.shallowClone(c.xla_HloInstructionProto, upb_arena, instruction); + } + } else return error.ComputationNotFound; + }; + + c.xla_HloInstructionProto_set_opcode(fused_root, upb.stringView("custom-call")); + c.xla_HloInstructionProto_set_custom_call_target(fused_root, upb.stringView("AwsNeuronNeff")); + c.xla_HloInstructionProto_set_backend_config(fused_root, blk: { + const neff_file = try std.fs.openFileAbsolute(neff_file_path, .{}); + defer neff_file.close(); + const stat = try neff_file.stat(); + const neff_buf = try allocator.alloc(u8, stat.size); + const size = try neff_file.readAll(neff_buf); + break :blk upb.stringView(neff_buf[0..size]); + }); + + const parameters_len = blk: { + var size: usize = undefined; + _ = c.xla_ProgramShapeProto_parameters( + c.xla_HloComputationProto_program_shape(entry), + &size, + ); + break :blk size; + }; + + { + var operand_ids: std.ArrayListUnmanaged(i64) = .initBuffer(c.xla_HloInstructionProto_resize_operand_ids(fused_root, parameters_len + 1, upb_arena)[0 .. parameters_len + 1]); + var new_instructions: std.ArrayListUnmanaged(*const c.xla_HloInstructionProto) = .initBuffer(@ptrCast(c.xla_HloComputationProto_resize_instructions(entry, parameters_len + 1, upb_arena)[0 .. parameters_len + 1])); + for (entry_instructions) |instruction| { + if (std.mem.eql(u8, upb.slice(c.xla_HloInstructionProto_opcode(instruction)) orelse continue, "parameter")) { + const id = c.xla_HloInstructionProto_id(instruction); + operand_ids.appendAssumeCapacity(id); + new_instructions.appendAssumeCapacity(instruction); + } + } + new_instructions.appendAssumeCapacity(fused_root); + } + + { + const fa = c.xla_HloInstructionProto_mutable_frontend_attributes(fused_root, upb_arena); + const map = c._xla_FrontendAttributes_map_mutable_upb_map(fa, upb_arena); + _ = c.upb_Map_Set( + map, + .{ .str_val = upb.stringView("valid_inputs") }, + .{ .str_val = blk: { + const valid_inputs_value = try allocator.alloc(u8, parameters_len * 2 - 1); + for (valid_inputs_value, 0..) |*char, i| { + char.* = if (i % 2 == 0) '1' else ','; + } + break :blk upb.stringView(valid_inputs_value); + } }, + upb_arena, + ); + } + + return try upb.serialize(hlo_module, upb_arena); +} + +fn neuronx_cc_(self: ?*c.PyObject, args_: [*c]*c.PyObject, nargs_: c.Py_ssize_t) !?*c.PyObject { + _ = self; + + var arena = std.heap.ArenaAllocator.init(std.heap.c_allocator); + defer arena.deinit(); + + const args = args_[0..@intCast(nargs_)]; + + const code = PyBytes_AsStringAndSize(args[0]); + const platform_version = PyBytes_AsStringAndSize(args[2]); + + const target = std.StaticStringMap([]const u8).initComptime(.{ + .{ "1.0", "inf1" }, + .{ "2.0", "trn1" }, + .{ "3.0", "trn2" }, + }).get(platform_version) orelse { + log.err("Unknown platform version: {s}\n", .{platform_version}); + return error.UnknownPlatformVersion; + }; + + var tmp_dir_buf: [std.fs.max_path_bytes]u8 = undefined; + const tmp_dir = try makeTempDir(&tmp_dir_buf, "zml-neuronxcc-"); + defer std.fs.deleteTreeAbsolute(tmp_dir) catch |err| { + log.err("Error deleting temporary directory {s}: {}\n", .{ tmp_dir, err }); + }; + + const code_file = try std.fs.path.join(arena.allocator(), &.{ tmp_dir, "file.code" }); + { + const file = try std.fs.cwd().createFile(code_file, .{ .truncate = true }); + defer file.close(); + try file.writeAll(code); + } + + const neff_file = try std.fs.path.join(arena.allocator(), &.{ tmp_dir, "file.neff" }); + + var neuronx_cc_buf: [std.fs.max_path_bytes]u8 = undefined; + var child = std.process.Child.init(&.{ + try stdx.fs.path.bufJoin(&neuronx_cc_buf, &.{ + stdx.fs.selfSharedObjectDirPath(), + "..", + "bin", + "neuronx-cc", + }), + "compile", + "--framework=XLA", + "--target", + target, + "--verbose=info", + "--enable-internal-neff-wrapper", + "--output", + neff_file, + "--optlevel=1", + // generic is the default, but it fails on transformers, force it + "--model-type=transformer", + // disable it, we do our own + "--auto-cast=none", + "--enable-fast-loading-neuron-binaries", + code_file, + }, arena.allocator()); + child.stdin_behavior = .Ignore; + child.stdout_behavior = .Inherit; + child.stderr_behavior = .Inherit; + child.cwd = tmp_dir; + _ = try child.spawnAndWait(); + + std.debug.print(">>>> {s}\n", .{tmp_dir}); + + const neff_hlo_bytes = wrapNeffAsCustomCall(arena.allocator(), code, neff_file) catch |err| { + log.err("Error wrapping NEFF as custom call: {}\n", .{err}); + return err; + }; + + return c.PyTuple_Pack( + 2, + c.PyLong_FromLongLong(0), + c.PyBytes_FromStringAndSize(@ptrCast(neff_hlo_bytes), @intCast(neff_hlo_bytes.len)), + ); +} + +fn neuronx_cc(self: ?*c.PyObject, args_: [*c]*c.PyObject, nargs_: c.Py_ssize_t) callconv(.c) ?*c.PyObject { + return neuronx_cc_(self, args_, nargs_) catch |err| { + log.err("Error in neuronx_cc: {}\n", .{err}); + + const none = c.Py_None(); + c.Py_IncRef(none); + const tuple = c.PyTuple_New(2) orelse { + c.Py_DecRef(none); + return null; + }; + _ = c.PyTuple_SetItem(tuple, 0, c.PyLong_FromLongLong(400).?); + _ = c.PyTuple_SetItem(tuple, 1, none); + return tuple; + }; +} + +pub export fn PyInit_libneuronxla() callconv(.c) ?*c.PyObject { + return c.PyModuleDef_Init(&module_def); +} diff --git a/runtimes/neuron/libpjrt_neuron.BUILD.bazel b/runtimes/neuron/libpjrt_neuron.BUILD.bazel index 5df16d6..01c0bed 100644 --- a/runtimes/neuron/libpjrt_neuron.BUILD.bazel +++ b/runtimes/neuron/libpjrt_neuron.BUILD.bazel @@ -4,18 +4,7 @@ load("@bazel_skylib//rules:copy_file.bzl", "copy_file") load("@bazel_skylib//rules:select_file.bzl", "select_file") load("@rules_cc//cc:cc_shared_library.bzl", "cc_shared_library") load("@zml//bazel:runfiles.bzl", "runfiles_to_default") - -cc_shared_library( - name = "zmlxneuron_so_", - shared_lib_name = "libzmlxneuron.so.0", - deps = ["@zml//runtimes/neuron:zmlxneuron_lib"], -) - -copy_file( - name = "zmlxneuron_so", - src = ":zmlxneuron_so_", - out = "lib/libzmlxneuron.so.0", -) +load("@rules_python//python:py_binary.bzl", "py_binary") runfiles_to_default( name = "libneuronxla_files", @@ -23,40 +12,33 @@ runfiles_to_default( ) select_file( - name = "libneuronpjrt_so", + name = "libneuronpjrt_so_orig", srcs = ":libneuronxla_files", subpath = "site-packages/libneuronxla/libneuronpjrt.so", ) patchelf( - name = "libneuronpjrt.patchelf", - add_needed = [ - "libpython3.11.so.1.0", - "libzmlxneuron.so.0", - "libnccom.so.2", - ], + name = "libneuronpjrt_so", set_rpath = '$ORIGIN', - shared_library = ":libneuronpjrt_so", + src = ":libneuronpjrt_so_orig", soname = "libneuronpjrt.so", ) patchelf( - name = "libpjrt_neuron_proxy.patchelf", + name = "libpjrt_neuron_so", set_rpath = '$ORIGIN', - add_needed = [ - "libz.so.1", - "libgomp.so.1", - ], - shared_library = "@zml//runtimes/neuron:libpjrt_neuron_proxy", + src = "@zml//runtimes/neuron:libpjrt_neuron", soname = "libpjrt_neuron.so", ) copy_to_directory( name = "sandbox", srcs = [ - ":zmlxneuron_so", - ":libneuronpjrt.patchelf", - ":libpjrt_neuron_proxy.patchelf", + ":libneuronpjrt_so", + ":libpjrt_neuron_so", + "@zml//runtimes/neuron:neuronx-cc_files", + "@zml//runtimes/neuron:libneuronxla", + "@rules_python//python:current_py_toolchain", "@aws-neuronx-runtime-lib//:libnrt.patchelf", "@aws-neuronx-runtime-lib//:libncfw.patchelf", "@aws-neuronx-collectives//:libnccom", @@ -64,21 +46,35 @@ copy_to_directory( "@libgomp1", ], replace_prefixes = { - "libneuronpjrt.patchelf": "lib", - "libpjrt_neuron_proxy.patchelf": "lib", + "runtimes/neuron/neuronx-cc": "bin/neuronx-cc", + "runtimes/neuron/libneuronxla.so": "site-packages/libneuronxla.so", + "libneuronpjrt_so": "lib", + "libpjrt_neuron_so": "lib", "libnrt.patchelf": "lib", "libncfw.patchelf": "lib", "lib/x86_64-linux-gnu": "lib", "usr/lib/x86_64-linux-gnu": "lib", "opt/neuron": "lib", }, - add_directory_to_runfiles = True, + exclude_srcs_patterns = [ + "**/tests/**", + "**/include/**", + "**/_solib_*/**", + ], + include_srcs_patterns = [ + "runtimes/neuron/neuronx-cc", + "**/lib**", + "lib/python*/**", + "site-packages/**", + ], + allow_overwrites = True, + add_directory_to_runfiles = False, include_external_repositories = ["**"], + hardlink = "on", ) cc_library( - name = "libpjrt_neuron", - data = [":sandbox"], + name = "libnrt_headers", deps = [ "@aws-neuronx-runtime-lib//:libnrt_headers", ], @@ -94,3 +90,9 @@ cc_library( ], visibility = ["@zml//runtimes/neuron:__subpackages__"], ) + +cc_library( + name = "libpjrt_neuron", + data = [":sandbox"], + visibility = ["@zml//runtimes/neuron:__subpackages__"], +) diff --git a/runtimes/neuron/libpjrt_neuron.zig b/runtimes/neuron/libpjrt_neuron.zig index 6485ecd..9a13d5e 100644 --- a/runtimes/neuron/libpjrt_neuron.zig +++ b/runtimes/neuron/libpjrt_neuron.zig @@ -1,39 +1,202 @@ const std = @import("std"); -const runfiles = @import("runfiles"); + const bazel_builtin = @import("bazel_builtin"); +const c = @import("c"); +const runfiles = @import("runfiles"); const stdx = @import("stdx"); const log = std.log.scoped(.@"zml/runtimes/neuron"); -pub export fn GetPjrtApi() *anyopaque { - var arena = std.heap.ArenaAllocator.init(std.heap.c_allocator); - defer arena.deinit(); - - var r_ = runfiles.Runfiles.create(.{ .allocator = arena.allocator() }) catch |err| { - stdx.debug.panic("Unable to find runfiles: {}", .{err}); - } orelse stdx.debug.panic("Runfiles not availeabwewefle", .{}); - - const source_repo = bazel_builtin.current_repository; - const r = r_.withSourceRepo(source_repo); - - var path_buf: [std.fs.max_path_bytes]u8 = undefined; - const sandbox_path = r.rlocation("libpjrt_neuron/sandbox", &path_buf) catch |err| { - stdx.debug.panic("Failed to find sandbox path for NEURON runtime: {}", .{err}); - } orelse stdx.debug.panic("No NEURON sandbox path found", .{}); - - var lib_path_buf: [std.fs.max_path_bytes]u8 = undefined; - const library = stdx.fs.path.bufJoinZ(&lib_path_buf, &.{ sandbox_path, "lib", "libneuronpjrt.so" }) catch unreachable; - - var lib: std.DynLib = blk: { - const handle = std.c.dlopen(library, .{ .LAZY = true, .GLOBAL = true, .NODELETE = true }) orelse { - stdx.debug.panic("Unable to dlopen plugin: {s}", .{library}); - }; - break :blk .{ .inner = .{ .handle = handle } }; - }; - - const sym = lib.lookup(*const fn () callconv(.C) *anyopaque, "GetPjrtApi") orelse { - stdx.debug.panic("Unable to find symbol GetPjrtApi in plugin: {s}", .{library}); - }; - - return sym(); +fn findFreeTcpPort() !u16 { + var address = std.net.Address.initIp4(.{ 127, 0, 0, 1 }, 0); + const sockfd = try std.posix.socket( + std.posix.AF.INET, + std.posix.SOCK.STREAM, + std.posix.IPPROTO.TCP, + ); + defer std.posix.close(sockfd); + var socklen = address.getOsSockLen(); + try std.posix.bind(sockfd, &address.any, socklen); + try std.posix.getsockname(sockfd, &address.any, &socklen); + return address.getPort(); +} + +pub export fn zmlxneuron_dlopen(filename: [*c]const u8, flags: c_int) ?*anyopaque { + const replacements: std.StaticStringMap([:0]const u8) = .initComptime(.{ + .{ "libnccom.so", "libnccom.so.2" }, + .{ "libnrt.so", "libnrt.so.1" }, + .{ "libncfw.so", "libncfw.so.2" }, + }); + + var buf: [std.fs.max_path_bytes]u8 = undefined; + const new_filename: [*c]const u8 = if (filename) |f| blk: { + const replacement = replacements.get(std.fs.path.basename(std.mem.span(f))) orelse break :blk f; + break :blk stdx.fs.path.bufJoinZ(&buf, &.{ + stdx.fs.selfSharedObjectDirPath(), + replacement, + }) catch unreachable; + } else null; + + return std.c.dlopen(new_filename, @bitCast(flags)); +} + +extern fn setenv(name: [*:0]const u8, value: [*:0]const u8, overwrite: c_int) c_int; +fn setupNeuronEnv() !void { + var buf: [256]u8 = undefined; + _ = setenv( + "NEURON_RT_ROOT_COMM_ID", + try std.fmt.bufPrintZ(&buf, "127.0.0.1:{d}", .{try findFreeTcpPort()}), + 1, + ); + _ = setenv( + "NEURON_INTERNAL_PJRT_C_API_VERSION", + std.fmt.comptimePrint("{d}.{d}", .{ + c.PJRT_API_MAJOR, + c.PJRT_API_MINOR, + }), + 1, + ); + _ = setenv( + "NEURON_RT_STOCHASTIC_ROUNDING_EN", + "1", + 1, + ); +} + +fn pyStatusCheck(status: c.PyStatus) void { + if (c.PyStatus_Exception(status) != 0) { + if (c.PyStatus_IsExit(status) != 0) { + std.process.exit(@intCast(status.exitcode)); + } + c.Py_ExitStatusException(status); + } +} + +fn toPosixPathW(file_path: []const u8) error{NameTooLong}![std.posix.PATH_MAX - 1:0]c.wchar_t { + if (file_path.len >= std.posix.PATH_MAX) return error.NameTooLong; + + var path_with_null: [std.posix.PATH_MAX - 1:0]c.wchar_t = undefined; + const len = c.mbstowcs(&path_with_null, file_path.ptr, file_path.len); + path_with_null[len] = 0; + return path_with_null; +} + +fn setupPythonEnv(sandbox_path: []const u8) !void { + const Static = struct { + var py_config: c.PyConfig = undefined; + }; + + { + var preconfig: c.PyPreConfig = undefined; + c.PyPreConfig_InitIsolatedConfig(&preconfig); + preconfig.utf8_mode = 1; + pyStatusCheck(c.Py_PreInitialize(&preconfig)); + } + + c.PyConfig_InitIsolatedConfig(&Static.py_config); + + Static.py_config.module_search_paths_set = 1; + Static.py_config.optimization_level = 2; + Static.py_config.write_bytecode = 0; + + { + var buf: [std.fs.max_path_bytes]u8 = undefined; + const home = try std.fmt.bufPrintZ(&buf, "{f}{d}.{d}", .{ + std.fs.path.fmtJoin(&.{ + sandbox_path, + "lib", + "python", + }), + c.PY_MAJOR_VERSION, + c.PY_MINOR_VERSION, + }); + pyStatusCheck(c.PyConfig_SetBytesString(&Static.py_config, &Static.py_config.home, home)); + pyStatusCheck(c.PyWideStringList_Append(&Static.py_config.module_search_paths, &try toPosixPathW(home))); + } + + { + var buf: [std.fs.max_path_bytes]u8 = undefined; + const site_packages = try stdx.fs.path.bufJoin(&buf, &.{ + sandbox_path, + "site-packages", + }); + pyStatusCheck(c.PyWideStringList_Append(&Static.py_config.module_search_paths, &try toPosixPathW(site_packages))); + } + + pyStatusCheck(c.Py_InitializeFromConfig(&Static.py_config)); + + // release the GIL + _ = c.PyEval_SaveThread(); +} + +// Duplicates a PJRT Api object while being careful about struct size differences +fn dupePjrtApi(api: *c.PJRT_Api) c.PJRT_Api { + var ret: c.PJRT_Api = undefined; + const struct_size = @min(@sizeOf(c.PJRT_Api), api.struct_size); + @memcpy( + std.mem.asBytes(&ret)[0..struct_size], + std.mem.asBytes(api)[0..struct_size], + ); + return ret; +} + +fn getPjrtApi() !*c.PJRT_Api { + const Static = struct { + var inner: *c.PJRT_Api = undefined; + var proxy: c.PJRT_Api = undefined; + }; + + var sandbox_path_buf: [std.fs.max_path_bytes]u8 = undefined; + const sandbox_path = try stdx.fs.path.bufJoin(&sandbox_path_buf, &.{ + stdx.fs.selfSharedObjectDirPath(), + "..", + }); + + try setupNeuronEnv(); + try setupPythonEnv(sandbox_path); + + Static.inner = blk: { + const GetPjrtApi_inner = GetPjrtApi_blk: { + var lib: std.DynLib = .{ + .inner = .{ + .handle = handle_blk: { + var lib_path_buf: [std.fs.max_path_bytes]u8 = undefined; + const library = try stdx.fs.path.bufJoinZ(&lib_path_buf, &.{ sandbox_path, "lib", "libneuronpjrt.so" }); + break :handle_blk std.c.dlopen(library, .{ .LAZY = true, .NODELETE = true }) orelse { + log.err("Unable to dlopen plugin: {?s}", .{std.mem.span(std.c.dlerror())}); + return error.DlOpenFailed; + }; + }, + }, + }; + + break :GetPjrtApi_blk lib.lookup(*const fn () callconv(.c) *c.PJRT_Api, "GetPjrtApi") orelse { + log.err("Unable to find symbol GetPjrtApi in plugin: {?s}", .{std.mem.span(std.c.dlerror())}); + return error.SymbolNotFound; + }; + }; + + break :blk GetPjrtApi_inner(); + }; + + Static.proxy = dupePjrtApi(Static.inner); + // Setup the API proxy functions + Static.proxy.PJRT_Plugin_Attributes = &struct { + const STRUCT_SIZE = 24; // according to the failing assertion + + fn call(args: [*c]c.PJRT_Plugin_Attributes_Args) callconv(.c) ?*c.PJRT_Error { + var new_args = args.*; + new_args.struct_size = @min(new_args.struct_size, STRUCT_SIZE); + return Static.inner.PJRT_Plugin_Attributes.?(&new_args); + } + }.call; + + return &Static.proxy; +} + +pub export fn GetPjrtApi() ?*c.PJRT_Api { + return getPjrtApi() catch |err| { + log.err("Failed to get PJRT API: {}", .{err}); + return null; + }; } diff --git a/runtimes/neuron/neuron.bzl b/runtimes/neuron/neuron.bzl index a277b35..e03e8ca 100644 --- a/runtimes/neuron/neuron.bzl +++ b/runtimes/neuron/neuron.bzl @@ -32,7 +32,7 @@ _NEURON_PACKAGES = { packages.patchelf( name = "libnrt.patchelf", src = "lib/libnrt.so.1", - set_rpath = '$ORIGIN', + set_rpath = "$ORIGIN", add_needed = [ # readelf -d ./opt/aws/neuron/libl/libncfw.so "libncfw.so.2", diff --git a/runtimes/neuron/neuron.zig b/runtimes/neuron/neuron.zig index fc97342..09d6a21 100644 --- a/runtimes/neuron/neuron.zig +++ b/runtimes/neuron/neuron.zig @@ -1,10 +1,9 @@ -const builtin = @import("builtin"); const std = @import("std"); +const builtin = @import("builtin"); const asynk = @import("async"); const bazel_builtin = @import("bazel_builtin"); const c = @import("c"); -const libneuronxla_pyenv = @import("libneuronxla_pyenv"); const pjrt = @import("pjrt"); const runfiles = @import("runfiles"); const stdx = @import("stdx"); @@ -26,99 +25,12 @@ fn isRunningOnEC2() !bool { var f = try asynk.File.open("/sys/devices/virtual/dmi/id/sys_vendor", .{ .mode = .read_only }); defer f.close() catch {}; - var buf = [_]u8{0} ** AmazonEC2.len; + var buf: [AmazonEC2.len]u8 = undefined; _ = try f.reader().readAll(&buf); return std.mem.eql(u8, &buf, AmazonEC2); } -fn toWchar(str: []const u8, out: [:0]c.wchar_t) [:0]c.wchar_t { - const len = c.mbstowcs(out.ptr, str.ptr, str.len); - out[len] = 0; - return out[0..len :0]; -} - -fn pyErrorOrExit(status: c.PyStatus) void { - if (c.PyStatus_Exception(status) != 0) { - if (c.PyStatus_IsExit(status) != 0) { - std.process.exit(@intCast(status.exitcode)); - } - c.Py_ExitStatusException(status); - } -} - -fn initialize(allocator: std.mem.Allocator, r_: *runfiles.Runfiles) !void { - { - var preconfig: c.PyPreConfig = undefined; - c.PyPreConfig_InitIsolatedConfig(&preconfig); - preconfig.utf8_mode = 1; - pyErrorOrExit(c.Py_PreInitialize(&preconfig)); - } - - var config: c.PyConfig = undefined; - c.PyConfig_InitIsolatedConfig(&config); - defer c.PyConfig_Clear(&config); - - const r = r_.withSourceRepo(bazel_builtin.current_repository); - - var buf: [std.fs.max_path_bytes]u8 = undefined; - var wbuf: [std.fs.max_path_bytes:0]c.wchar_t = undefined; - - { - const path = (try r.rlocation(libneuronxla_pyenv.home, &buf)).?; - const wpath = toWchar(std.fs.path.dirname(path).?, &wbuf); - pyErrorOrExit(c.PyConfig_SetString(&config, &config.home, wpath.ptr)); - } - - { - config.module_search_paths_set = 1; - for (libneuronxla_pyenv.modules) |module| { - const path = (try r.rlocation(module, &buf)).?; - const wline = toWchar(std.fs.path.dirname(path).?, &wbuf); - pyErrorOrExit(c.PyWideStringList_Append(&config.module_search_paths, wline.ptr)); - } - } - - { - const neuronx_cc = (try r.rlocation("zml/runtimes/neuron/neuronx-cc/neuronx-cc", &buf)).?; - const neuronx_cc_path = std.fs.path.dirname(neuronx_cc).?; - const path = std.posix.getenv("PATH") orelse ""; - const new_path = try std.fmt.allocPrintZ(allocator, "{s}:{s}", .{ neuronx_cc_path, path }); - _ = c.setenv("PATH", new_path.ptr, 1); - } - - pyErrorOrExit(c.Py_InitializeFromConfig(&config)); - - // release the GIL - _ = c.PyEval_SaveThread(); -} - -fn comptimeStrJoin(comptime separator: [:0]const u8, comptime slices: []const [:0]const u8) [:0]const u8 { - comptime var ret = slices[0]; - inline for (slices[1..]) |slice| { - ret = ret ++ separator ++ slice; - } - return ret; -} - -pub fn setNeuronCCFlags() void { - // See neuronxcc reference: - // https://awsdocs-neuron.readthedocs-hosted.com/en/latest/compiler/neuronx-cc/api-reference-guide/neuron-compiler-cli-reference-guide.html#neuron-compiler-cli-reference-guide - _ = c.setenv("NEURON_CC_FLAGS", comptimeStrJoin(" ", &.{ - // 30% faster, no visible speed difference on llama - "--optlevel=1", - // generic is the default, but it fails on transformers, force it - "--model-type=transformer", - // disable it, we do our own - "--auto-cast=none", - "--enable-fast-loading-neuron-binaries", - }), 1); - - // Enable stochastic rounding - // https://awsdocs-neuron.readthedocs-hosted.com/en/latest/general/arch/neuron-features/rounding-modes.html - _ = c.setenv("NEURON_RT_STOCHASTIC_ROUNDING_EN", "1", 1); -} - pub fn load() !*const pjrt.Api { if (comptime !isEnabled()) { return error.Unavailable; @@ -143,15 +55,12 @@ pub fn load() !*const pjrt.Api { const source_repo = bazel_builtin.current_repository; const r = r_.withSourceRepo(source_repo); - var path_buf: [std.fs.max_path_bytes]u8 = undefined; - const sandbox_path = try r.rlocation("libpjrt_neuron/sandbox", &path_buf) orelse { + var sandbox_path_buf: [std.fs.max_path_bytes]u8 = undefined; + const sandbox_path = try r.rlocation("libpjrt_neuron/sandbox", &sandbox_path_buf) orelse { log.err("Failed to find sandbox path for NEURON runtime", .{}); return error.FileNotFound; }; - setNeuronCCFlags(); - try initialize(arena.allocator(), &r_); - return blk: { var lib_path_buf: [std.fs.max_path_bytes]u8 = undefined; const path = try stdx.fs.path.bufJoinZ(&lib_path_buf, &.{ sandbox_path, "lib", "libpjrt_neuron.so" }); diff --git a/runtimes/neuron/neuronx-cc.zig b/runtimes/neuron/neuronx-cc.zig new file mode 100644 index 0000000..828a1d4 --- /dev/null +++ b/runtimes/neuron/neuronx-cc.zig @@ -0,0 +1,89 @@ +const std = @import("std"); + +const c = @import("c"); +const runfiles = @import("runfiles"); + +fn pyStatusCheck(status: c.PyStatus) void { + if (c.PyStatus_Exception(status) != 0) { + if (c.PyStatus_IsExit(status) != 0) { + std.process.exit(@intCast(status.exitcode)); + } + c.Py_ExitStatusException(status); + } +} + +pub fn toPosixPathW(file_path: []const u8) error{NameTooLong}![std.posix.PATH_MAX - 1:0]c.wchar_t { + if (file_path.len >= std.posix.PATH_MAX) return error.NameTooLong; + + var path_with_null: [std.posix.PATH_MAX - 1:0]c.wchar_t = undefined; + const len = c.mbstowcs(&path_with_null, file_path.ptr, file_path.len); + path_with_null[len] = 0; + return path_with_null; +} + +pub fn main() !void { + var arena = std.heap.ArenaAllocator.init(std.heap.c_allocator); + defer arena.deinit(); + + { + var preconfig: c.PyPreConfig = undefined; + c.PyPreConfig_InitIsolatedConfig(&preconfig); + preconfig.utf8_mode = 1; + pyStatusCheck(c.Py_PreInitialize(&preconfig)); + } + + var config: c.PyConfig = undefined; + c.PyConfig_InitIsolatedConfig(&config); + defer c.PyConfig_Clear(&config); + + config.module_search_paths_set = 1; + config.optimization_level = 2; + config.write_bytecode = 0; + + _ = c.PyConfig_SetBytesArgv(&config, @intCast(std.os.argv.len), @ptrCast(std.os.argv)); + + var self_exe_dir_buf: [std.fs.max_path_bytes]u8 = undefined; + const self_exe_dir = try std.fs.selfExeDirPath(&self_exe_dir_buf); + + { + var buf: [std.fs.max_path_bytes]u8 = undefined; + const home = try std.fmt.bufPrintZ(&buf, "{f}{d}.{d}", .{ + std.fs.path.fmtJoin(&.{ + self_exe_dir, + "..", + "lib", + "python", + }), + c.PY_MAJOR_VERSION, + c.PY_MINOR_VERSION, + }); + pyStatusCheck(c.PyConfig_SetBytesString(&config, &config.home, home)); + pyStatusCheck(c.PyWideStringList_Append(&config.module_search_paths, &try toPosixPathW(home))); + } + + { + var buf: [std.fs.max_path_bytes]u8 = undefined; + const site_packages = try std.fmt.bufPrint(&buf, "{f}", .{ + std.fs.path.fmtJoin(&.{ + self_exe_dir, + "..", + "site-packages", + }), + }); + pyStatusCheck(c.PyWideStringList_Append(&config.module_search_paths, &try toPosixPathW(site_packages))); + } + + pyStatusCheck(c.Py_InitializeFromConfig(&config)); + defer c.Py_Finalize(); + + const neuronxcc_main = blk: { + const module = c.PyImport_ImportModule("neuronxcc.driver.CommandDriver"); + std.debug.print(">>> MODULE: {any}\n", .{module}); + defer c.Py_DecRef(module); + break :blk c.PyObject_GetAttrString(module, "main"); + }; + defer c.Py_DecRef(neuronxcc_main); + + const result = c.PyObject_CallNoArgs(neuronxcc_main); + defer c.Py_DecRef(result); +} diff --git a/runtimes/neuron/libpjrt_neuron.h b/runtimes/neuron/nrt.h similarity index 100% rename from runtimes/neuron/libpjrt_neuron.h rename to runtimes/neuron/nrt.h diff --git a/runtimes/neuron/pyenv.bzl b/runtimes/neuron/pyenv.bzl deleted file mode 100644 index 11c27db..0000000 --- a/runtimes/neuron/pyenv.bzl +++ /dev/null @@ -1,41 +0,0 @@ -load("@bazel_skylib//lib:paths.bzl", "paths") -load("@rules_python//python:defs.bzl", "PyInfo") - -def runfile_path(ctx, runfile): - return paths.normalize(ctx.workspace_name + "/" + runfile.short_path) - -def _pyenv_zig_impl(ctx): - py_toolchain = ctx.toolchains["@rules_python//python:toolchain_type"].py3_runtime - content = "pub const home: [:0]const u8 = {};\n".format(repr(runfile_path(ctx, py_toolchain.interpreter))) - - modules = [] - for file in py_toolchain.files.to_list(): - if file.basename == "__hello__.py": - modules.append(repr(runfile_path(ctx, file))) - - imports = depset([], transitive = [ - dep[PyInfo].imports - for dep in ctx.attr.deps - ]) - modules.extend([ - repr("{}/__init__.py".format(imp)) - for imp in imports.to_list() - ]) - content += "pub const modules: []const [:0]const u8 = &.{{ {} }};\n".format(", ".join(modules)) - - f = ctx.actions.declare_file("{}.zig".format(ctx.label.name)) - ctx.actions.write(f, content) - - return [ - DefaultInfo( - files = depset([f]), - ), - ] - -pyenv_zig = rule( - implementation = _pyenv_zig_impl, - attrs = { - "deps": attr.label_list(providers = [PyInfo]), - }, - toolchains = ["@rules_python//python:toolchain_type"], -) diff --git a/runtimes/neuron/requirements.lock.txt b/runtimes/neuron/requirements.lock.txt index c5994c0..9332b2f 100644 --- a/runtimes/neuron/requirements.lock.txt +++ b/runtimes/neuron/requirements.lock.txt @@ -1,121 +1,108 @@ # This file was autogenerated by uv via the following command: -# bazel run @@//runtimes/neuron:update_requirements +# bazel run //runtimes/neuron:requirements.update --index-url https://pypi.org/simple --extra-index-url https://pip.repos.neuron.amazonaws.com --find-links https://mirror.zml.ai/pypi/aws-neuronx-runtime-discovery/index.html -boto3==1.39.9 \ - --hash=sha256:5bc85e9fdec4e21ef5ca2c22b4d51a3e32b53f3da36ce51f5a3ea4dbde07b132 \ - --hash=sha256:e3d3a6b617e1575e7ec854c820a882ab2e189a0421e74dc0dca2c9e13d4370a5 +boto3==1.40.14 \ + --hash=sha256:ab5798a03582d09c0de132d080c9aee53d5647b6461261a5b7621170ec80d92b \ + --hash=sha256:d1d9998fc2b9619fc796c859d263ac81793d783e79331be62931b353dd1b68b9 # via libneuronxla -botocore==1.39.9 \ - --hash=sha256:02f141c2849e4589a79feea245ce4ecc478d48b7865572445af8aae3b041772d \ - --hash=sha256:a9691cbe03a3bc8b2720b3c36e5c5a2eecace6acd72bfb1107f00e75edaec4f3 +botocore==1.40.14 \ + --hash=sha256:06c852be83543c8d45e18a530abcad31659db8b16bcb66fd371c930d38017e7c \ + --hash=sha256:1810494c8c4190f20f9e17f2da4f7ee91eda863f429c6d690ea17ec41a8f83c4 # via # boto3 # libneuronxla # s3transfer -certifi==2025.7.14 \ - --hash=sha256:6b31f564a415d79ee77df69d757bb49a5bb53bd9f756cbbe24394ffd6fc1f4b2 \ - --hash=sha256:8ea99dbdfaaf2ba2f9bac77b9249ef62ec5218e7c2b2e903378ed5fccf765995 +certifi==2025.8.3 \ + --hash=sha256:e564105f78ded564e3ae7c923924435e1daa7463faeab5bb932bc53ffae63407 \ + --hash=sha256:f6c12493cfb1b06ba2ff328595af9350c65d6644968e5d3a2ffd78699af217a5 # via requests -charset-normalizer==3.4.2 \ - --hash=sha256:005fa3432484527f9732ebd315da8da8001593e2cf46a3d817669f062c3d9ed4 \ - --hash=sha256:046595208aae0120559a67693ecc65dd75d46f7bf687f159127046628178dc45 \ - --hash=sha256:0c29de6a1a95f24b9a1aa7aefd27d2487263f00dfd55a77719b530788f75cff7 \ - --hash=sha256:0c8c57f84ccfc871a48a47321cfa49ae1df56cd1d965a09abe84066f6853b9c0 \ - --hash=sha256:0f5d9ed7f254402c9e7d35d2f5972c9bbea9040e99cd2861bd77dc68263277c7 \ - --hash=sha256:18dd2e350387c87dabe711b86f83c9c78af772c748904d372ade190b5c7c9d4d \ - --hash=sha256:1b1bde144d98e446b056ef98e59c256e9294f6b74d7af6846bf5ffdafd687a7d \ - --hash=sha256:1c95a1e2902a8b722868587c0e1184ad5c55631de5afc0eb96bc4b0d738092c0 \ - --hash=sha256:1cad5f45b3146325bb38d6855642f6fd609c3f7cad4dbaf75549bf3b904d3184 \ - --hash=sha256:21b2899062867b0e1fde9b724f8aecb1af14f2778d69aacd1a5a1853a597a5db \ - --hash=sha256:24498ba8ed6c2e0b56d4acbf83f2d989720a93b41d712ebd4f4979660db4417b \ - --hash=sha256:25a23ea5c7edc53e0f29bae2c44fcb5a1aa10591aae107f2a2b2583a9c5cbc64 \ - --hash=sha256:289200a18fa698949d2b39c671c2cc7a24d44096784e76614899a7ccf2574b7b \ - --hash=sha256:28a1005facc94196e1fb3e82a3d442a9d9110b8434fc1ded7a24a2983c9888d8 \ - --hash=sha256:32fc0341d72e0f73f80acb0a2c94216bd704f4f0bce10aedea38f30502b271ff \ - --hash=sha256:36b31da18b8890a76ec181c3cf44326bf2c48e36d393ca1b72b3f484113ea344 \ - --hash=sha256:3c21d4fca343c805a52c0c78edc01e3477f6dd1ad7c47653241cf2a206d4fc58 \ - --hash=sha256:3fddb7e2c84ac87ac3a947cb4e66d143ca5863ef48e4a5ecb83bd48619e4634e \ - --hash=sha256:43e0933a0eff183ee85833f341ec567c0980dae57c464d8a508e1b2ceb336471 \ - --hash=sha256:4a476b06fbcf359ad25d34a057b7219281286ae2477cc5ff5e3f70a246971148 \ - --hash=sha256:4e594135de17ab3866138f496755f302b72157d115086d100c3f19370839dd3a \ - --hash=sha256:50bf98d5e563b83cc29471fa114366e6806bc06bc7a25fd59641e41445327836 \ - --hash=sha256:5a9979887252a82fefd3d3ed2a8e3b937a7a809f65dcb1e068b090e165bbe99e \ - --hash=sha256:5baececa9ecba31eff645232d59845c07aa030f0c81ee70184a90d35099a0e63 \ - --hash=sha256:5bf4545e3b962767e5c06fe1738f951f77d27967cb2caa64c28be7c4563e162c \ - --hash=sha256:6333b3aa5a12c26b2a4d4e7335a28f1475e0e5e17d69d55141ee3cab736f66d1 \ - --hash=sha256:65c981bdbd3f57670af8b59777cbfae75364b483fa8a9f420f08094531d54a01 \ - --hash=sha256:68a328e5f55ec37c57f19ebb1fdc56a248db2e3e9ad769919a58672958e8f366 \ - --hash=sha256:6a0289e4589e8bdfef02a80478f1dfcb14f0ab696b5a00e1f4b8a14a307a3c58 \ - --hash=sha256:6b66f92b17849b85cad91259efc341dce9c1af48e2173bf38a85c6329f1033e5 \ - --hash=sha256:6c9379d65defcab82d07b2a9dfbfc2e95bc8fe0ebb1b176a3190230a3ef0e07c \ - --hash=sha256:6fc1f5b51fa4cecaa18f2bd7a003f3dd039dd615cd69a2afd6d3b19aed6775f2 \ - --hash=sha256:70f7172939fdf8790425ba31915bfbe8335030f05b9913d7ae00a87d4395620a \ - --hash=sha256:721c76e84fe669be19c5791da68232ca2e05ba5185575086e384352e2c309597 \ - --hash=sha256:7222ffd5e4de8e57e03ce2cef95a4c43c98fcb72ad86909abdfc2c17d227fc1b \ - --hash=sha256:75d10d37a47afee94919c4fab4c22b9bc2a8bf7d4f46f87363bcf0573f3ff4f5 \ - --hash=sha256:76af085e67e56c8816c3ccf256ebd136def2ed9654525348cfa744b6802b69eb \ - --hash=sha256:770cab594ecf99ae64c236bc9ee3439c3f46be49796e265ce0cc8bc17b10294f \ - --hash=sha256:7a6ab32f7210554a96cd9e33abe3ddd86732beeafc7a28e9955cdf22ffadbab0 \ - --hash=sha256:7c48ed483eb946e6c04ccbe02c6b4d1d48e51944b6db70f697e089c193404941 \ - --hash=sha256:7f56930ab0abd1c45cd15be65cc741c28b1c9a34876ce8c17a2fa107810c0af0 \ - --hash=sha256:8075c35cd58273fee266c58c0c9b670947c19df5fb98e7b66710e04ad4e9ff86 \ - --hash=sha256:8272b73e1c5603666618805fe821edba66892e2870058c94c53147602eab29c7 \ - --hash=sha256:82d8fd25b7f4675d0c47cf95b594d4e7b158aca33b76aa63d07186e13c0e0ab7 \ - --hash=sha256:844da2b5728b5ce0e32d863af26f32b5ce61bc4273a9c720a9f3aa9df73b1455 \ - --hash=sha256:8755483f3c00d6c9a77f490c17e6ab0c8729e39e6390328e42521ef175380ae6 \ - --hash=sha256:915f3849a011c1f593ab99092f3cecfcb4d65d8feb4a64cf1bf2d22074dc0ec4 \ - --hash=sha256:926ca93accd5d36ccdabd803392ddc3e03e6d4cd1cf17deff3b989ab8e9dbcf0 \ - --hash=sha256:982bb1e8b4ffda883b3d0a521e23abcd6fd17418f6d2c4118d257a10199c0ce3 \ - --hash=sha256:98f862da73774290f251b9df8d11161b6cf25b599a66baf087c1ffe340e9bfd1 \ - --hash=sha256:9cbfacf36cb0ec2897ce0ebc5d08ca44213af24265bd56eca54bee7923c48fd6 \ - --hash=sha256:a370b3e078e418187da8c3674eddb9d983ec09445c99a3a263c2011993522981 \ - --hash=sha256:a955b438e62efdf7e0b7b52a64dc5c3396e2634baa62471768a64bc2adb73d5c \ - --hash=sha256:aa6af9e7d59f9c12b33ae4e9450619cf2488e2bbe9b44030905877f0b2324980 \ - --hash=sha256:aa88ca0b1932e93f2d961bf3addbb2db902198dca337d88c89e1559e066e7645 \ - --hash=sha256:aaeeb6a479c7667fbe1099af9617c83aaca22182d6cf8c53966491a0f1b7ffb7 \ - --hash=sha256:aaf27faa992bfee0264dc1f03f4c75e9fcdda66a519db6b957a3f826e285cf12 \ - --hash=sha256:b2680962a4848b3c4f155dc2ee64505a9c57186d0d56b43123b17ca3de18f0fa \ - --hash=sha256:b2d318c11350e10662026ad0eb71bb51c7812fc8590825304ae0bdd4ac283acd \ - --hash=sha256:b33de11b92e9f75a2b545d6e9b6f37e398d86c3e9e9653c4864eb7e89c5773ef \ - --hash=sha256:b3daeac64d5b371dea99714f08ffc2c208522ec6b06fbc7866a450dd446f5c0f \ - --hash=sha256:be1e352acbe3c78727a16a455126d9ff83ea2dfdcbc83148d2982305a04714c2 \ - --hash=sha256:bee093bf902e1d8fc0ac143c88902c3dfc8941f7ea1d6a8dd2bcb786d33db03d \ - --hash=sha256:c72fbbe68c6f32f251bdc08b8611c7b3060612236e960ef848e0a517ddbe76c5 \ - --hash=sha256:c9e36a97bee9b86ef9a1cf7bb96747eb7a15c2f22bdb5b516434b00f2a599f02 \ - --hash=sha256:cddf7bd982eaa998934a91f69d182aec997c6c468898efe6679af88283b498d3 \ - --hash=sha256:cf713fe9a71ef6fd5adf7a79670135081cd4431c2943864757f0fa3a65b1fafd \ - --hash=sha256:d11b54acf878eef558599658b0ffca78138c8c3655cf4f3a4a673c437e67732e \ - --hash=sha256:d41c4d287cfc69060fa91cae9683eacffad989f1a10811995fa309df656ec214 \ - --hash=sha256:d524ba3f1581b35c03cb42beebab4a13e6cdad7b36246bd22541fa585a56cccd \ - --hash=sha256:daac4765328a919a805fa5e2720f3e94767abd632ae410a9062dff5412bae65a \ - --hash=sha256:db4c7bf0e07fc3b7d89ac2a5880a6a8062056801b83ff56d8464b70f65482b6c \ - --hash=sha256:dc7039885fa1baf9be153a0626e337aa7ec8bf96b0128605fb0d77788ddc1681 \ - --hash=sha256:dccab8d5fa1ef9bfba0590ecf4d46df048d18ffe3eec01eeb73a42e0d9e7a8ba \ - --hash=sha256:dedb8adb91d11846ee08bec4c8236c8549ac721c245678282dcb06b221aab59f \ - --hash=sha256:e45ba65510e2647721e35323d6ef54c7974959f6081b58d4ef5d87c60c84919a \ - --hash=sha256:e53efc7c7cee4c1e70661e2e112ca46a575f90ed9ae3fef200f2a25e954f4b28 \ - --hash=sha256:e635b87f01ebc977342e2697d05b56632f5f879a4f15955dfe8cef2448b51691 \ - --hash=sha256:e70e990b2137b29dc5564715de1e12701815dacc1d056308e2b17e9095372a82 \ - --hash=sha256:e8082b26888e2f8b36a042a58307d5b917ef2b1cacab921ad3323ef91901c71a \ - --hash=sha256:e8323a9b031aa0393768b87f04b4164a40037fb2a3c11ac06a03ffecd3618027 \ - --hash=sha256:e92fca20c46e9f5e1bb485887d074918b13543b1c2a1185e69bb8d17ab6236a7 \ - --hash=sha256:eb30abc20df9ab0814b5a2524f23d75dcf83cde762c161917a2b4b7b55b1e518 \ - --hash=sha256:eba9904b0f38a143592d9fc0e19e2df0fa2e41c3c3745554761c5f6447eedabf \ - --hash=sha256:ef8de666d6179b009dce7bcb2ad4c4a779f113f12caf8dc77f0162c29d20490b \ - --hash=sha256:efd387a49825780ff861998cd959767800d54f8308936b21025326de4b5a42b9 \ - --hash=sha256:f0aa37f3c979cf2546b73e8222bbfa3dc07a641585340179d768068e3455e544 \ - --hash=sha256:f4074c5a429281bf056ddd4c5d3b740ebca4d43ffffe2ef4bf4d2d05114299da \ - --hash=sha256:f69a27e45c43520f5487f27627059b64aaf160415589230992cec34c5e18a509 \ - --hash=sha256:fb707f3e15060adf5b7ada797624a6c6e0138e2a26baa089df64c68ee98e040f \ - --hash=sha256:fcbe676a55d7445b22c10967bceaaf0ee69407fbe0ece4d032b6eb8d4565982a \ - --hash=sha256:fdb20a30fe1175ecabed17cbf7812f7b804b8a315a25f24678bcdf120a90077f +charset-normalizer==3.4.3 \ + --hash=sha256:00237675befef519d9af72169d8604a067d92755e84fe76492fef5441db05b91 \ + --hash=sha256:02425242e96bcf29a49711b0ca9f37e451da7c70562bc10e8ed992a5a7a25cc0 \ + --hash=sha256:027b776c26d38b7f15b26a5da1044f376455fb3766df8fc38563b4efbc515154 \ + --hash=sha256:07a0eae9e2787b586e129fdcbe1af6997f8d0e5abaa0bc98c0e20e124d67e601 \ + --hash=sha256:0cacf8f7297b0c4fcb74227692ca46b4a5852f8f4f24b3c766dd94a1075c4884 \ + --hash=sha256:0e78314bdc32fa80696f72fa16dc61168fda4d6a0c014e0380f9d02f0e5d8a07 \ + --hash=sha256:0f2be7e0cf7754b9a30eb01f4295cc3d4358a479843b31f328afd210e2c7598c \ + --hash=sha256:13faeacfe61784e2559e690fc53fa4c5ae97c6fcedb8eb6fb8d0a15b475d2c64 \ + --hash=sha256:14c2a87c65b351109f6abfc424cab3927b3bdece6f706e4d12faaf3d52ee5efe \ + --hash=sha256:1606f4a55c0fd363d754049cdf400175ee96c992b1f8018b993941f221221c5f \ + --hash=sha256:16a8770207946ac75703458e2c743631c79c59c5890c80011d536248f8eaa432 \ + --hash=sha256:18343b2d246dc6761a249ba1fb13f9ee9a2bcd95decc767319506056ea4ad4dc \ + --hash=sha256:18b97b8404387b96cdbd30ad660f6407799126d26a39ca65729162fd810a99aa \ + --hash=sha256:1bb60174149316da1c35fa5233681f7c0f9f514509b8e399ab70fea5f17e45c9 \ + --hash=sha256:1e8ac75d72fa3775e0b7cb7e4629cec13b7514d928d15ef8ea06bca03ef01cae \ + --hash=sha256:1ef99f0456d3d46a50945c98de1774da86f8e992ab5c77865ea8b8195341fc19 \ + --hash=sha256:2001a39612b241dae17b4687898843f254f8748b796a2e16f1051a17078d991d \ + --hash=sha256:23b6b24d74478dc833444cbd927c338349d6ae852ba53a0d02a2de1fce45b96e \ + --hash=sha256:252098c8c7a873e17dd696ed98bbe91dbacd571da4b87df3736768efa7a792e4 \ + --hash=sha256:257f26fed7d7ff59921b78244f3cd93ed2af1800ff048c33f624c87475819dd7 \ + --hash=sha256:2c322db9c8c89009a990ef07c3bcc9f011a3269bc06782f916cd3d9eed7c9312 \ + --hash=sha256:30a96e1e1f865f78b030d65241c1ee850cdf422d869e9028e2fc1d5e4db73b92 \ + --hash=sha256:30d006f98569de3459c2fc1f2acde170b7b2bd265dc1943e87e1a4efe1b67c31 \ + --hash=sha256:31a9a6f775f9bcd865d88ee350f0ffb0e25936a7f930ca98995c05abf1faf21c \ + --hash=sha256:320e8e66157cc4e247d9ddca8e21f427efc7a04bbd0ac8a9faf56583fa543f9f \ + --hash=sha256:34a7f768e3f985abdb42841e20e17b330ad3aaf4bb7e7aeeb73db2e70f077b99 \ + --hash=sha256:3653fad4fe3ed447a596ae8638b437f827234f01a8cd801842e43f3d0a6b281b \ + --hash=sha256:3cd35b7e8aedeb9e34c41385fda4f73ba609e561faedfae0a9e75e44ac558a15 \ + --hash=sha256:3cfb2aad70f2c6debfbcb717f23b7eb55febc0bb23dcffc0f076009da10c6392 \ + --hash=sha256:416175faf02e4b0810f1f38bcb54682878a4af94059a1cd63b8747244420801f \ + --hash=sha256:41d1fc408ff5fdfb910200ec0e74abc40387bccb3252f3f27c0676731df2b2c8 \ + --hash=sha256:42e5088973e56e31e4fa58eb6bd709e42fc03799c11c42929592889a2e54c491 \ + --hash=sha256:4ca4c094de7771a98d7fbd67d9e5dbf1eb73efa4f744a730437d8a3a5cf994f0 \ + --hash=sha256:511729f456829ef86ac41ca78c63a5cb55240ed23b4b737faca0eb1abb1c41bc \ + --hash=sha256:53cd68b185d98dde4ad8990e56a58dea83a4162161b1ea9272e5c9182ce415e0 \ + --hash=sha256:585f3b2a80fbd26b048a0be90c5aae8f06605d3c92615911c3a2b03a8a3b796f \ + --hash=sha256:5b413b0b1bfd94dbf4023ad6945889f374cd24e3f62de58d6bb102c4d9ae534a \ + --hash=sha256:5d8d01eac18c423815ed4f4a2ec3b439d654e55ee4ad610e153cf02faf67ea40 \ + --hash=sha256:6aab0f181c486f973bc7262a97f5aca3ee7e1437011ef0c2ec04b5a11d16c927 \ + --hash=sha256:6cf8fd4c04756b6b60146d98cd8a77d0cdae0e1ca20329da2ac85eed779b6849 \ + --hash=sha256:6fb70de56f1859a3f71261cbe41005f56a7842cc348d3aeb26237560bfa5e0ce \ + --hash=sha256:6fce4b8500244f6fcb71465d4a4930d132ba9ab8e71a7859e6a5d59851068d14 \ + --hash=sha256:70bfc5f2c318afece2f5838ea5e4c3febada0be750fcf4775641052bbba14d05 \ + --hash=sha256:73dc19b562516fc9bcf6e5d6e596df0b4eb98d87e4f79f3ae71840e6ed21361c \ + --hash=sha256:74d77e25adda8581ffc1c720f1c81ca082921329452eba58b16233ab1842141c \ + --hash=sha256:78deba4d8f9590fe4dae384aeff04082510a709957e968753ff3c48399f6f92a \ + --hash=sha256:86df271bf921c2ee3818f0522e9a5b8092ca2ad8b065ece5d7d9d0e9f4849bcc \ + --hash=sha256:88ab34806dea0671532d3f82d82b85e8fc23d7b2dd12fa837978dad9bb392a34 \ + --hash=sha256:8999f965f922ae054125286faf9f11bc6932184b93011d138925a1773830bbe9 \ + --hash=sha256:8dcfc373f888e4fb39a7bc57e93e3b845e7f462dacc008d9749568b1c4ece096 \ + --hash=sha256:939578d9d8fd4299220161fdd76e86c6a251987476f5243e8864a7844476ba14 \ + --hash=sha256:96b2b3d1a83ad55310de8c7b4a2d04d9277d5591f40761274856635acc5fcb30 \ + --hash=sha256:a2d08ac246bb48479170408d6c19f6385fa743e7157d716e144cad849b2dd94b \ + --hash=sha256:b256ee2e749283ef3ddcff51a675ff43798d92d746d1a6e4631bf8c707d22d0b \ + --hash=sha256:b5e3b2d152e74e100a9e9573837aba24aab611d39428ded46f4e4022ea7d1942 \ + --hash=sha256:b89bc04de1d83006373429975f8ef9e7932534b8cc9ca582e4db7d20d91816db \ + --hash=sha256:bd28b817ea8c70215401f657edef3a8aa83c29d447fb0b622c35403780ba11d5 \ + --hash=sha256:c60e092517a73c632ec38e290eba714e9627abe9d301c8c8a12ec32c314a2a4b \ + --hash=sha256:c6dbd0ccdda3a2ba7c2ecd9d77b37f3b5831687d8dc1b6ca5f56a4880cc7b7ce \ + --hash=sha256:c6e490913a46fa054e03699c70019ab869e990270597018cef1d8562132c2669 \ + --hash=sha256:c6f162aabe9a91a309510d74eeb6507fab5fff92337a15acbe77753d88d9dcf0 \ + --hash=sha256:c6fd51128a41297f5409deab284fecbe5305ebd7e5a1f959bee1c054622b7018 \ + --hash=sha256:cc34f233c9e71701040d772aa7490318673aa7164a0efe3172b2981218c26d93 \ + --hash=sha256:cc9370a2da1ac13f0153780040f465839e6cccb4a1e44810124b4e22483c93fe \ + --hash=sha256:ccf600859c183d70eb47e05a44cd80a4ce77394d1ac0f79dbd2dd90a69a3a049 \ + --hash=sha256:ce571ab16d890d23b5c278547ba694193a45011ff86a9162a71307ed9f86759a \ + --hash=sha256:cf1ebb7d78e1ad8ec2a8c4732c7be2e736f6e5123a4146c5b89c9d1f585f8cef \ + --hash=sha256:d0e909868420b7049dafd3a31d45125b31143eec59235311fc4c57ea26a4acd2 \ + --hash=sha256:d22dbedd33326a4a5190dd4fe9e9e693ef12160c77382d9e87919bce54f3d4ca \ + --hash=sha256:d716a916938e03231e86e43782ca7878fb602a125a91e7acb8b5112e2e96ac16 \ + --hash=sha256:d79c198e27580c8e958906f803e63cddb77653731be08851c7df0b1a14a8fc0f \ + --hash=sha256:d95bfb53c211b57198bb91c46dd5a2d8018b3af446583aab40074bf7988401cb \ + --hash=sha256:e28e334d3ff134e88989d90ba04b47d84382a828c061d0d1027b1b12a62b39b1 \ + --hash=sha256:ec557499516fc90fd374bf2e32349a2887a876fbf162c160e3c01b6849eaf557 \ + --hash=sha256:fb6fecfd65564f208cbf0fba07f107fb661bcd1a7c389edbced3f7a493f70e37 \ + --hash=sha256:fb731e5deb0c7ef82d698b0f4c5bb724633ee2a489401594c5c88b02e6cb15f7 \ + --hash=sha256:fb7f67a1bfa6e40b438170ebdc8158b78dc465a5a67b6dde178a46987b244a72 \ + --hash=sha256:fd10de089bcdcd1be95a2f73dbe6254798ec1bda9f450d5828c96f93e2536b9c \ + --hash=sha256:fdabf8315679312cfa71302f9bd509ded4f2f263fb5b765cf1433b39106c3cc9 # via requests -ec2-metadata==2.14.0 \ - --hash=sha256:b2f83381722072efa28f65dc37e726c0abd3a0532f227b3fa6a2ad68460c7feb \ - --hash=sha256:fc4f9e893da1c049761221ec3d2208cce4872953b8960e1ed1f7120ff4ca3512 +ec2-metadata==2.15.0 \ + --hash=sha256:b066b268091ca40369fdd644b90a34ecbfd78225bce6f0e269f0bf9ce3ab1b75 \ + --hash=sha256:daeec6b93229623942ac61f5ccc3026d0079ea1e7c84a5b18a3fb631db85600f # via neuronx-cc idna==3.10 \ --hash=sha256:12f65c9b470abda6dc35cf8e63cc574b1c52b11df2c86030af0ac09b01b13ea9 \ @@ -169,45 +156,56 @@ jmespath==1.0.1 \ # via # boto3 # botocore -libneuronxla==2.2.4410.0+835a67fb \ - --hash=sha256:37e5f08482ef1a9a844c2894de6d91c03400a092764e54ab67612456be439f16 +libneuronxla==2.2.8201.0+f46ac1ef \ + --hash=sha256:c5bb9d2c4ab45a786561e09919ef6a3e3b7839bb7b18f539130e8c8659a29b1e # via -r runtimes/neuron/requirements.in lockfile==0.12.2 \ --hash=sha256:6aed02de03cba24efabcd600b30540140634fc06cfa603822d508d5361e9f799 \ --hash=sha256:6c3cb24f344923d30b2785d5ad75182c8ea7ac1b6171b08657258ec7429d50fa # via python-daemon -ml-dtypes==0.5.1 \ - --hash=sha256:023ce2f502efd4d6c1e0472cc58ce3640d051d40e71e27386bed33901e201327 \ - --hash=sha256:05f23447a1c20ddf4dc7c2c661aa9ed93fcb2658f1017c204d1e758714dc28a8 \ - --hash=sha256:12651420130ee7cc13059fc56dac6ad300c3af3848b802d475148c9defd27c23 \ - --hash=sha256:141b2ea2f20bb10802ddca55d91fe21231ef49715cfc971998e8f2a9838f3dbe \ - --hash=sha256:15ad0f3b0323ce96c24637a88a6f44f6713c64032f27277b069f285c3cf66478 \ - --hash=sha256:1b7fbe5571fdf28fd3aaab3ef4aafc847de9ebf263be959958c1ca58ec8eadf5 \ - --hash=sha256:26ebcc69d7b779c8f129393e99732961b5cc33fcff84090451f448c89b0e01b4 \ - --hash=sha256:6f462f5eca22fb66d7ff9c4744a3db4463af06c49816c4b6ac89b16bfcdc592e \ - --hash=sha256:6f76232163b5b9c34291b54621ee60417601e2e4802a188a0ea7157cd9b323f4 \ - --hash=sha256:7000b6e4d8ef07542c05044ec5d8bbae1df083b3f56822c3da63993a113e716f \ - --hash=sha256:810512e2eccdfc3b41eefa3a27402371a3411453a1efc7e9c000318196140fed \ - --hash=sha256:8f2c028954f16ede77902b223a8da2d9cbb3892375b85809a5c3cfb1587960c4 \ - --hash=sha256:9626d0bca1fb387d5791ca36bacbba298c5ef554747b7ebeafefb4564fc83566 \ - --hash=sha256:ac5b58559bb84a95848ed6984eb8013249f90b6bab62aa5acbad876e256002c9 \ - --hash=sha256:ad4953c5eb9c25a56d11a913c2011d7e580a435ef5145f804d98efa14477d390 \ - --hash=sha256:aefedc579ece2f8fb38f876aa7698204ee4c372d0e54f1c1ffa8ca580b54cc60 \ - --hash=sha256:afb2009ac98da274e893e03162f6269398b2b00d947e7057ee2469a921d58135 \ - --hash=sha256:b8a9d46b4df5ae2135a8e8e72b465448ebbc1559997f4f9304a9ecc3413efb5b \ - --hash=sha256:bd73f51957949069573ff783563486339a9285d72e2f36c18e0c1aa9ca7eb190 \ - --hash=sha256:bf9975bda82a99dc935f2ae4c83846d86df8fd6ba179614acac8e686910851da \ - --hash=sha256:c09526488c3a9e8b7a23a388d4974b670a9a3dd40c5c8a61db5593ce9b725bab \ - --hash=sha256:c9945669d3dadf8acb40ec2e57d38c985d8c285ea73af57fc5b09872c516106d \ - --hash=sha256:d13755f8e8445b3870114e5b6240facaa7cb0c3361e54beba3e07fa912a6e12b \ - --hash=sha256:fd918d4e6a4e0c110e2e05be7a7814d10dc1b95872accbf6512b80a109b71ae1 +ml-dtypes==0.5.3 \ + --hash=sha256:01de48de4537dc3c46e684b969a40ec36594e7eeb7c69e9a093e7239f030a28a \ + --hash=sha256:0a1d68a7cb53e3f640b2b6a34d12c0542da3dd935e560fdf463c0c77f339fc20 \ + --hash=sha256:0cd5a6c711b5350f3cbc2ac28def81cd1c580075ccb7955e61e9d8f4bfd40d24 \ + --hash=sha256:0e44a3761f64bc009d71ddb6d6c71008ba21b53ab6ee588dadab65e2fa79eafc \ + --hash=sha256:156418abeeda48ea4797db6776db3c5bdab9ac7be197c1233771e0880c304057 \ + --hash=sha256:19f6c3a4f635c2fc9e2aa7d91416bd7a3d649b48350c51f7f715a09370a90d93 \ + --hash=sha256:1b255acada256d1fa8c35ed07b5f6d18bc21d1556f842fbc2d5718aea2cd9e55 \ + --hash=sha256:1db60c154989af253f6c4a34e8a540c2c9dce4d770784d426945e09908fbb177 \ + --hash=sha256:2db74788fc01914a3c7f7da0763427280adfc9cd377e9604b6b64eb8097284bd \ + --hash=sha256:4a177b882667c69422402df6ed5c3428ce07ac2c1f844d8a1314944651439458 \ + --hash=sha256:4cae435a68861660af81fa3c5af16b70ca11a17275c5b662d9c6f58294e0f113 \ + --hash=sha256:5103856a225465371fe119f2fef737402b705b810bd95ad5f348e6e1a6ae21af \ + --hash=sha256:58e39349d820b5702bb6f94ea0cb2dc8ec62ee81c0267d9622067d8333596a46 \ + --hash=sha256:5ab039ffb40f3dc0aeeeba84fd6c3452781b5e15bef72e2d10bcb33e4bbffc39 \ + --hash=sha256:5ee72568d46b9533ad54f78b1e1f3067c0534c5065120ea8ecc6f210d22748b3 \ + --hash=sha256:66c2756ae6cfd7f5224e355c893cfd617fa2f747b8bbd8996152cbdebad9a184 \ + --hash=sha256:6936283b56d74fbec431ca57ce58a90a908fdbd14d4e2d22eea6d72bb208a7b7 \ + --hash=sha256:8b1a6e231b0770f2894910f1dce6d2f31d65884dbf7668f9b08d73623cdca909 \ + --hash=sha256:8bb9cd1ce63096567f5f42851f5843b5a0ea11511e50039a7649619abfb4ba6d \ + --hash=sha256:93c36a08a6d158db44f2eb9ce3258e53f24a9a4a695325a689494f0fdbc71770 \ + --hash=sha256:95ce33057ba4d05df50b1f3cfefab22e351868a843b3b15a46c65836283670c9 \ + --hash=sha256:9849ce7267444c0a717c80c6900997de4f36e2815ce34ac560a3edb2d9a64cd2 \ + --hash=sha256:9d55ea7f7baf2aed61bf1872116cefc9d0c3693b45cae3916897ee27ef4b835e \ + --hash=sha256:a4f39b9bf6555fab9bfb536cf5fdd1c1c727e8d22312078702e9ff005354b37f \ + --hash=sha256:aec640bd94c4c85c0d11e2733bd13cbb10438fb004852996ec0efbc6cacdaf70 \ + --hash=sha256:aecbd7c5272c82e54d5b99d8435fd10915d1bc704b7df15e4d9ca8dc3902be61 \ + --hash=sha256:bda32ce212baa724e03c68771e5c69f39e584ea426bfe1a701cb01508ffc7035 \ + --hash=sha256:bdcf26c2dbc926b8a35ec8cbfad7eff1a8bd8239e12478caca83a1fc2c400dc2 \ + --hash=sha256:bdf40d2aaabd3913dec11840f0d0ebb1b93134f99af6a0a4fd88ffe924928ab4 \ + --hash=sha256:c205cac07d24a29840c163d6469f61069ce4b065518519216297fc2f261f8db9 \ + --hash=sha256:c3f5ae0309d9f888fd825c2e9d0241102fadaca81d888f26f845bc8c13c1e4ee \ + --hash=sha256:cd7c0bb22d4ff86d65ad61b5dd246812e8993fbc95b558553624c33e8b6903ea \ + --hash=sha256:d0f730a17cf4f343b2c7ad50cee3bd19e969e793d2be6ed911f43086460096e4 \ + --hash=sha256:da65e5fd3eea434ccb8984c3624bc234ddcc0d9f4c81864af611aaebcc08a50e \ + --hash=sha256:e12e29764a0e66a7a31e9b8bf1de5cc0423ea72979f45909acd4292de834ccd3 # via neuronx-cc networkx==2.8.8 \ --hash=sha256:230d388117af870fce5647a3c52401fcf753e94720e6ea6b4197a5355648885e \ --hash=sha256:e435dfa75b1d7195c7b8378c3859f0445cd88c6b0375c181ed66823a9ceb7524 # via neuronx-cc -neuronx-cc==2.19.8089.0+8ab9f450 \ - --hash=sha256:fbb05d8a0da8f54954c03ff2bea659f627309110c00efd54f523d877ddd6b88d +neuronx-cc==2.20.9961.0+0acef03a \ + --hash=sha256:22db1e09975f400b74edc2fab24a1b86d967386f060f5a29a3ec21756ecf5b0c # via # -r runtimes/neuron/requirements.in # libneuronxla @@ -256,16 +254,16 @@ pgzip==0.3.5 \ --hash=sha256:4e13ab66ecface5c51c5af51d8cd676aa51675cf85df000f501a86cf38c208c1 \ --hash=sha256:dd35510f59f6bd6b64e31c4baf90c10cdbb2775235fcc079b14b404fbd7f46bf # via neuronx-cc -protobuf==6.31.1 \ - --hash=sha256:0414e3aa5a5f3ff423828e1e6a6e907d6c65c1d5b7e6e975793d5590bdeecc16 \ - --hash=sha256:426f59d2964864a1a366254fa703b8632dcec0790d8862d30034d8245e1cd447 \ - --hash=sha256:4ee898bf66f7a8b0bd21bce523814e6fbd8c6add948045ce958b73af7e8878c6 \ - --hash=sha256:6f1227473dc43d44ed644425268eb7c2e488ae245d51c6866d19fe158e207402 \ - --hash=sha256:720a6c7e6b77288b85063569baae8536671b39f15cc22037ec7045658d80489e \ - --hash=sha256:7fa17d5a29c2e04b7d90e5e32388b8bfd0e7107cd8e616feef7ed3fa6bdab5c9 \ - --hash=sha256:8764cf4587791e7564051b35524b72844f845ad0bb011704c3736cce762d8fe9 \ - --hash=sha256:a40fc12b84c154884d7d4c4ebd675d5b3b5283e155f324049ae396b95ddebc39 \ - --hash=sha256:d8cac4c982f0b957a4dc73a80e2ea24fab08e679c0de9deb835f4a12d69aca9a +protobuf==6.32.0 \ + --hash=sha256:15eba1b86f193a407607112ceb9ea0ba9569aed24f93333fe9a497cf2fda37d3 \ + --hash=sha256:501fe6372fd1c8ea2a30b4d9be8f87955a64d6be9c88a973996cef5ef6f0abf1 \ + --hash=sha256:75a2aab2bd1aeb1f5dc7c5f33bcb11d82ea8c055c9becbb41c26a8c43fd7092c \ + --hash=sha256:7db8ed09024f115ac877a1427557b838705359f047b2ff2f2b2364892d19dacb \ + --hash=sha256:84f9e3c1ff6fb0308dbacb0950d8aa90694b0d0ee68e75719cb044b7078fe741 \ + --hash=sha256:a81439049127067fc49ec1d36e25c6ee1d1a2b7be930675f919258d03c04e7d2 \ + --hash=sha256:a8bdbb2f009cfc22a36d031f22a625a38b615b5e19e558a7b756b3279723e68e \ + --hash=sha256:ba377e5b67b908c8f3072a57b63e2c6a4cbd18aea4ed98d2584350dbf46f2783 \ + --hash=sha256:d52691e5bee6c860fff9a1c86ad26a13afbeb4b168cd4445c922b7e2cf85aaf0 # via neuronx-cc psutil==7.0.0 \ --hash=sha256:101d71dc322e3cffd7cea0650b09b3d08b8e7c4109dd6809fe452dfd00e58b25 \ @@ -287,9 +285,9 @@ python-dateutil==2.9.0.post0 \ --hash=sha256:37dd54208da7e1cd875388217d5e00ebd4179249f90fb72437e91a35459a0ad3 \ --hash=sha256:a8b2bc7bffae282281c8140a97d3aa9c14da0b136dfe83f850eea9a5f7470427 # via botocore -requests==2.32.4 \ - --hash=sha256:27babd3cda2a6d50b30443204ee89830707d396671944c998b5975b031ac2b2c \ - --hash=sha256:27d0316682c8a29834d3264820024b62a36942083d52caf2f14c0591336d3422 +requests==2.32.5 \ + --hash=sha256:2462f94637a34fd532264295e186976db0f5d453d1cdd31473c85a6a161affb6 \ + --hash=sha256:dbba0bac56e100853db0ea71b82b4dfd5fe2bf6d3754a8893c3af500cec7d7cf # via # ec2-metadata # requests-unixsocket diff --git a/runtimes/neuron/zmlxneuron.c b/runtimes/neuron/zmlxneuron.c deleted file mode 100644 index a5740be..0000000 --- a/runtimes/neuron/zmlxneuron.c +++ /dev/null @@ -1,24 +0,0 @@ -#include -#include - -void *zmlxneuron_dlopen(const char *filename, int flags) -{ - if (filename != NULL) - { - char *replacements[] = { - "libnccom.so", "libnccom.so.2", - "libnrt.so", "libnrt.so.1", - "libncfw.so", "libncfw.so.2", - NULL, NULL, - }; - for (int i = 0; replacements[i] != NULL; i += 2) - { - if (strcmp(filename, replacements[i]) == 0) - { - filename = replacements[i + 1]; - break; - } - } - } - return dlopen(filename, flags); -} diff --git a/upb/BUILD.bazel b/upb/BUILD.bazel index ca8ebc3..316cad3 100644 --- a/upb/BUILD.bazel +++ b/upb/BUILD.bazel @@ -1,12 +1,16 @@ load("@rules_zig//zig:defs.bzl", "zig_library") cc_library( - name = "empty", + name = "upb_c", + hdrs = ["upb.h"], + deps = [ + "@com_google_protobuf//upb/message:copy", + ], ) zig_library( name = "upb", main = "upb.zig", - deps = [":empty"], + deps = [":upb_c"], visibility = ["//visibility:public"], ) diff --git a/upb/upb.h b/upb/upb.h new file mode 100644 index 0000000..dbec357 --- /dev/null +++ b/upb/upb.h @@ -0,0 +1 @@ +#include "upb/message/copy.h" diff --git a/upb/upb.zig b/upb/upb.zig index 8470285..27d1502 100644 --- a/upb/upb.zig +++ b/upb/upb.zig @@ -38,6 +38,14 @@ pub const ParseError = error{ Unknown, } || std.mem.Allocator.Error; +pub fn deepClone(comptime UpbType: type, arena: *c.upb_Arena, msg: *const UpbType) std.mem.Allocator.Error!*UpbType { + return @ptrCast(c.upb_Message_DeepClone(@ptrCast(msg), Minitable(UpbType), arena) orelse return std.mem.Allocator.Error.OutOfMemory); +} + +pub fn shallowClone(comptime UpbType: type, arena: *c.upb_Arena, msg: *const UpbType) std.mem.Allocator.Error!*UpbType { + return @ptrCast(c.upb_Message_ShallowClone(@ptrCast(msg), Minitable(UpbType), arena) orelse return std.mem.Allocator.Error.OutOfMemory); +} + pub fn stringView(data: ?[]const u8) c.upb_StringView { return if (data) |d| c.upb_StringView_FromDataAndSize(d.ptr, d.len) else .{}; } @@ -100,7 +108,7 @@ pub fn parseEx(comptime UpbType: type, arena: *c.upb_Arena, data: []const u8, op } pub fn parse(comptime UpbType: type, arena: *c.upb_Arena, data: []const u8) ParseError!*UpbType { - return parseEx(arena, data, .{}); + return parseEx(UpbType, arena, data, .{}); } pub fn new(comptime UpbType: type, arena: *c.upb_Arena) std.mem.Allocator.Error!*UpbType {