| (function (global, factory) { |
| typeof exports === 'object' && typeof module !== 'undefined' ? factory(exports, require('perf_hooks'), require('ws')) : |
| typeof define === 'function' && define.amd ? define(['exports', 'perf_hooks', 'ws'], factory) : |
| (global = typeof globalThis !== 'undefined' ? globalThis : global || self, factory(global.tvmjs = {}, global.perf_hooks, global.ws)); |
| })(this, (function (exports, perf_hooks, ws) { 'use strict'; |
|
|
| function _interopDefaultLegacy (e) { return e && typeof e === 'object' && 'default' in e ? e : { 'default': e }; } |
|
|
| var perf_hooks__default = _interopDefaultLegacy(perf_hooks); |
| var ws__default = _interopDefaultLegacy(ws); |
|
|
| var commonjsGlobal = typeof globalThis !== 'undefined' ? globalThis : typeof window !== 'undefined' ? window : typeof global !== 'undefined' ? global : typeof self !== 'undefined' ? self : {}; |
|
|
| function unwrapExports (x) { |
| return x && x.__esModule && Object.prototype.hasOwnProperty.call(x, 'default') ? x['default'] : x; |
| } |
|
|
| function createCommonjsModule(fn, module) { |
| return module = { exports: {} }, fn(module, module.exports), module.exports; |
| } |
|
|
| var support = createCommonjsModule(function (module, exports) { |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| Object.defineProperty(exports, "__esModule", { value: true }); |
| exports.wasmPath = exports.assert = exports.Uint8ArrayToString = exports.StringToUint8Array = void 0; |
| |
| |
| |
| |
| |
| function StringToUint8Array(str) { |
| const arr = new Uint8Array(str.length + 1); |
| for (let i = 0; i < str.length; ++i) { |
| arr[i] = str.charCodeAt(i); |
| } |
| arr[str.length] = 0; |
| return arr; |
| } |
| exports.StringToUint8Array = StringToUint8Array; |
| |
| |
| |
| |
| |
| function Uint8ArrayToString(arr) { |
| const ret = []; |
| for (const ch of arr) { |
| ret.push(String.fromCharCode(ch)); |
| } |
| return ret.join(""); |
| } |
| exports.Uint8ArrayToString = Uint8ArrayToString; |
| |
| |
| |
| |
| |
| function assert(condition, msg) { |
| if (!condition) { |
| throw new Error("AssertError:" + (msg || "")); |
| } |
| } |
| exports.assert = assert; |
| |
| |
| |
| |
| function wasmPath() { |
| return __dirname + "/wasm"; |
| } |
| exports.wasmPath = wasmPath; |
|
|
| }); |
|
|
| unwrapExports(support); |
| support.wasmPath; |
| support.assert; |
| support.Uint8ArrayToString; |
| support.StringToUint8Array; |
|
|
| var memory = createCommonjsModule(function (module, exports) { |
| Object.defineProperty(exports, "__esModule", { value: true }); |
| exports.CachedCallStack = exports.Memory = void 0; |
|
|
| |
| |
| |
| class Memory { |
| constructor(memory) { |
| this.wasm32 = true; |
| this.memory = memory; |
| this.buffer = this.memory.buffer; |
| this.viewU8 = new Uint8Array(this.buffer); |
| this.viewU16 = new Uint16Array(this.buffer); |
| this.viewI32 = new Int32Array(this.buffer); |
| this.viewU32 = new Uint32Array(this.buffer); |
| this.viewF32 = new Float32Array(this.buffer); |
| this.viewF64 = new Float64Array(this.buffer); |
| } |
| loadU8(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| return this.viewU8[ptr >> 0]; |
| } |
| loadU16(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| return this.viewU16[ptr >> 1]; |
| } |
| loadU32(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| return this.viewU32[ptr >> 2]; |
| } |
| loadI32(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| return this.viewI32[ptr >> 2]; |
| } |
| loadI64(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| const base = ptr >> 2; |
| |
| return this.viewI32[base]; |
| } |
| loadF32(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| return this.viewF32[ptr >> 2]; |
| } |
| loadF64(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| return this.viewF64[ptr >> 3]; |
| } |
| loadPointer(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| if (this.wasm32) { |
| return this.loadU32(ptr); |
| } |
| else { |
| return this.loadI64(ptr); |
| } |
| } |
| loadUSize(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| if (this.wasm32) { |
| return this.loadU32(ptr); |
| } |
| else { |
| return this.loadI64(ptr); |
| } |
| } |
| sizeofPtr() { |
| return this.wasm32 ? 4 : 8 ; |
| } |
| |
| |
| |
| |
| |
| loadRawBytes(ptr, numBytes) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| const result = new Uint8Array(numBytes); |
| result.set(this.viewU8.slice(ptr, ptr + numBytes)); |
| return result; |
| } |
| |
| |
| |
| |
| |
| loadTVMBytes(ptr) { |
| const data = this.loadPointer(ptr); |
| const length = this.loadUSize(ptr + this.sizeofPtr()); |
| return this.loadRawBytes(data, length); |
| } |
| |
| |
| |
| |
| loadCString(ptr) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| |
| const ret = []; |
| let ch = 1; |
| while (ch != 0) { |
| ch = this.viewU8[ptr]; |
| if (ch != 0) { |
| ret.push(String.fromCharCode(ch)); |
| } |
| ++ptr; |
| } |
| return ret.join(""); |
| } |
| |
| |
| |
| |
| |
| storeRawBytes(ptr, bytes) { |
| if (this.buffer != this.memory.buffer) { |
| this.updateViews(); |
| } |
| this.viewU8.set(bytes, ptr); |
| } |
| |
| |
| |
| updateViews() { |
| this.buffer = this.memory.buffer; |
| this.viewU8 = new Uint8Array(this.buffer); |
| this.viewU16 = new Uint16Array(this.buffer); |
| this.viewI32 = new Int32Array(this.buffer); |
| this.viewU32 = new Uint32Array(this.buffer); |
| this.viewF32 = new Float32Array(this.buffer); |
| this.viewF64 = new Float64Array(this.buffer); |
| } |
| } |
| exports.Memory = Memory; |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| class CachedCallStack { |
| constructor(memory, allocSpace, freeSpace) { |
| |
| this.tempArgs = []; |
| this.stackTop = 0; |
| this.basePtr = 0; |
| this.addressToSetTargetValue = []; |
| const initCallStackSize = 128; |
| this.memory = memory; |
| this.cAllocSpace = allocSpace; |
| this.cFreeSpace = freeSpace; |
| this.buffer = new ArrayBuffer(initCallStackSize); |
| this.basePtr = this.cAllocSpace(initCallStackSize); |
| this.viewU8 = new Uint8Array(this.buffer); |
| this.viewI32 = new Int32Array(this.buffer); |
| this.viewU32 = new Uint32Array(this.buffer); |
| this.viewF64 = new Float64Array(this.buffer); |
| this.updateViews(); |
| } |
| dispose() { |
| if (this.basePtr != 0) { |
| this.cFreeSpace(this.basePtr); |
| this.basePtr = 0; |
| } |
| } |
| |
| |
| |
| reset() { |
| this.stackTop = 0; |
| support.assert(this.addressToSetTargetValue.length == 0); |
| while (this.tempArgs.length != 0) { |
| this.tempArgs.pop().dispose(); |
| } |
| } |
| |
| |
| |
| |
| |
| |
| |
| commitToWasmMemory(nbytes = this.stackTop) { |
| |
| while (this.addressToSetTargetValue.length != 0) { |
| const [targetOffset, valueOffset] = this.addressToSetTargetValue.pop(); |
| this.storePtr(targetOffset, this.ptrFromOffset(valueOffset)); |
| } |
| this.memory.storeRawBytes(this.basePtr, this.viewU8.slice(0, nbytes)); |
| } |
| |
| |
| |
| |
| |
| allocRawBytes(nbytes) { |
| |
| nbytes = ((nbytes + 7) >> 3) << 3; |
| if (this.stackTop + nbytes > this.buffer.byteLength) { |
| const newSize = Math.max(this.buffer.byteLength * 2, this.stackTop + nbytes); |
| const oldU8 = this.viewU8; |
| this.buffer = new ArrayBuffer(newSize); |
| this.updateViews(); |
| this.viewU8.set(oldU8); |
| if (this.basePtr != 0) { |
| this.cFreeSpace(this.basePtr); |
| } |
| this.basePtr = this.cAllocSpace(newSize); |
| } |
| const retOffset = this.stackTop; |
| this.stackTop += nbytes; |
| return retOffset; |
| } |
| |
| |
| |
| |
| |
| allocPtrArray(count) { |
| return this.allocRawBytes(this.memory.sizeofPtr() * count); |
| } |
| |
| |
| |
| |
| |
| ptrFromOffset(offset) { |
| return this.basePtr + offset; |
| } |
| |
| storePtr(offset, value) { |
| if (this.memory.wasm32) { |
| this.storeU32(offset, value); |
| } |
| else { |
| this.storeI64(offset, value); |
| } |
| } |
| storeUSize(offset, value) { |
| if (this.memory.wasm32) { |
| this.storeU32(offset, value); |
| } |
| else { |
| this.storeI64(offset, value); |
| } |
| } |
| storeI32(offset, value) { |
| this.viewI32[offset >> 2] = value; |
| } |
| storeU32(offset, value) { |
| this.viewU32[offset >> 2] = value; |
| } |
| storeI64(offset, value) { |
| |
| |
| const low = value & 0xffffffff; |
| const base = offset >> 2; |
| this.viewI32[base] = low; |
| this.viewI32[base + 1] = 0; |
| } |
| storeF64(offset, value) { |
| this.viewF64[offset >> 3] = value; |
| } |
| storeRawBytes(offset, bytes) { |
| this.viewU8.set(bytes, offset); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| allocThenSetArgString(offset, data) { |
| const strOffset = this.allocRawBytes(data.length + 1); |
| this.storeRawBytes(strOffset, support.StringToUint8Array(data)); |
| this.addressToSetTargetValue.push([offset, strOffset]); |
| } |
| |
| |
| |
| |
| |
| |
| |
| allocThenSetArgBytes(offset, data) { |
| |
| const headerOffset = this.allocRawBytes(this.memory.sizeofPtr() * 2); |
| const dataOffset = this.allocRawBytes(data.length); |
| this.storeRawBytes(dataOffset, data); |
| this.storeUSize(headerOffset + this.memory.sizeofPtr(), data.length); |
| this.addressToSetTargetValue.push([offset, headerOffset]); |
| this.addressToSetTargetValue.push([headerOffset, dataOffset]); |
| } |
| |
| |
| |
| updateViews() { |
| this.viewU8 = new Uint8Array(this.buffer); |
| this.viewI32 = new Int32Array(this.buffer); |
| this.viewU32 = new Uint32Array(this.buffer); |
| this.viewF64 = new Float64Array(this.buffer); |
| } |
| } |
| exports.CachedCallStack = CachedCallStack; |
|
|
| }); |
|
|
| unwrapExports(memory); |
| memory.CachedCallStack; |
| memory.Memory; |
|
|
| var environment = createCommonjsModule(function (module, exports) { |
| Object.defineProperty(exports, "__esModule", { value: true }); |
| exports.Environment = void 0; |
|
|
| |
| |
| |
| |
| |
| function detectLibraryProvider(importObject) { |
| if (importObject["wasmLibraryProvider"] && |
| importObject["wasmLibraryProvider"]["start"] && |
| importObject["wasmLibraryProvider"]["imports"] !== undefined) { |
| const item = importObject; |
| |
| return { |
| imports: item.wasmLibraryProvider.imports, |
| start: (inst) => { |
| item.wasmLibraryProvider.start(inst); |
| }, |
| }; |
| } |
| else if (importObject["imports"] && importObject["start"] !== undefined) { |
| return importObject; |
| } |
| else if (importObject["wasiImport"] && importObject["start"] !== undefined) { |
| |
| return { |
| imports: { |
| "wasi_snapshot_preview1": importObject["wasiImport"], |
| }, |
| start: (inst) => { |
| importObject["start"](inst); |
| } |
| }; |
| } |
| else { |
| return undefined; |
| } |
| } |
| |
| |
| |
| class Environment { |
| constructor(importObject = {}, logger = console.log) { |
| |
| |
| |
| |
| |
| |
| |
| this.packedCFuncTable = [ |
| undefined, |
| ]; |
| |
| |
| |
| this.packedCFuncTableFreeId = []; |
| this.logger = logger; |
| this.libProvider = detectLibraryProvider(importObject); |
| |
| if (this.libProvider !== undefined) { |
| this.imports = this.libProvider.imports; |
| } |
| else { |
| this.imports = importObject; |
| } |
| |
| this.imports.env = this.environment(this.imports.env); |
| } |
| |
| start(inst) { |
| if (this.libProvider !== undefined) { |
| this.libProvider.start(inst); |
| } |
| } |
| environment(initEnv) { |
| |
| const defaultEnv = { |
| "__cxa_thread_atexit": () => { }, |
| |
| "emscripten_notify_memory_growth": (index) => { } |
| }; |
| const wasmPackedCFunc = (args, typeCodes, nargs, ret, resourceHandle) => { |
| const cfunc = this.packedCFuncTable[resourceHandle]; |
| support.assert(cfunc !== undefined); |
| return cfunc(args, typeCodes, nargs, ret, resourceHandle); |
| }; |
| const wasmPackedCFuncFinalizer = (resourceHandle) => { |
| this.packedCFuncTable[resourceHandle] = undefined; |
| this.packedCFuncTableFreeId.push(resourceHandle); |
| }; |
| const newEnv = { |
| TVMWasmPackedCFunc: wasmPackedCFunc, |
| TVMWasmPackedCFuncFinalizer: wasmPackedCFuncFinalizer, |
| "__console_log": (msg) => { |
| this.logger(msg); |
| } |
| }; |
| return Object.assign(defaultEnv, initEnv, newEnv); |
| } |
| } |
| exports.Environment = Environment; |
|
|
| }); |
|
|
| unwrapExports(environment); |
| environment.Environment; |
|
|
| var webgpu = createCommonjsModule(function (module, exports) { |
| var __awaiter = (commonjsGlobal && commonjsGlobal.__awaiter) || function (thisArg, _arguments, P, generator) { |
| function adopt(value) { return value instanceof P ? value : new P(function (resolve) { resolve(value); }); } |
| return new (P || (P = Promise))(function (resolve, reject) { |
| function fulfilled(value) { try { step(generator.next(value)); } catch (e) { reject(e); } } |
| function rejected(value) { try { step(generator["throw"](value)); } catch (e) { reject(e); } } |
| function step(result) { result.done ? resolve(result.value) : adopt(result.value).then(fulfilled, rejected); } |
| step((generator = generator.apply(thisArg, _arguments || [])).next()); |
| }); |
| }; |
| Object.defineProperty(exports, "__esModule", { value: true }); |
| exports.WebGPUContext = exports.detectGPUDevice = void 0; |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
|
|
| |
| |
| |
| function detectGPUDevice() { |
| return __awaiter(this, void 0, void 0, function* () { |
| if (typeof navigator !== "undefined" && navigator.gpu !== undefined) { |
| const adapter = yield navigator.gpu.requestAdapter({ "powerPreference": "high-performance" }); |
| if (adapter == null) { |
| throw Error("Cannot find adapter that matches the request"); |
| } |
| const computeMB = (value) => { |
| return Math.ceil(value / (1 << 20)) + "MB"; |
| }; |
| |
| const requiedMaxBufferSize = 1 << 30; |
| if (requiedMaxBufferSize > adapter.limits.maxBufferSize) { |
| throw Error(`Cannot initialize runtime because of requested maxBufferSize ` + |
| `exceeds limit. requested=${computeMB(requiedMaxBufferSize)}, ` + |
| `limit=${computeMB(adapter.limits.maxBufferSize)}. ` + |
| `This error may be caused by an older version of the browser (e.g. Chrome 112). ` + |
| `You can try to upgrade your browser to Chrome 113 or later.`); |
| } |
| const requiredMaxStorageBufferBindingSize = 1 << 30; |
| if (requiredMaxStorageBufferBindingSize > adapter.limits.maxStorageBufferBindingSize) { |
| throw Error(`Cannot initialize runtime because of requested maxStorageBufferBindingSize ` + |
| `exceeds limit. requested=${computeMB(requiredMaxStorageBufferBindingSize)}, ` + |
| `limit=${computeMB(adapter.limits.maxStorageBufferBindingSize)}. `); |
| } |
| const requiredMaxComputeWorkgroupStorageSize = 32 << 10; |
| if (requiredMaxComputeWorkgroupStorageSize > adapter.limits.maxComputeWorkgroupStorageSize) { |
| throw Error(`Cannot initialize runtime because of requested maxComputeWorkgroupStorageSize ` + |
| `exceeds limit. requested=${requiredMaxComputeWorkgroupStorageSize}, ` + |
| `limit=${adapter.limits.maxComputeWorkgroupStorageSize}. `); |
| } |
| const adapterInfo = yield adapter.requestAdapterInfo(); |
| const device = yield adapter.requestDevice({ |
| requiredLimits: { |
| maxBufferSize: requiedMaxBufferSize, |
| maxStorageBufferBindingSize: requiredMaxStorageBufferBindingSize, |
| maxComputeWorkgroupStorageSize: requiredMaxComputeWorkgroupStorageSize, |
| } |
| }); |
| return { |
| adapter: adapter, |
| adapterInfo: adapterInfo, |
| device: device |
| }; |
| } |
| else { |
| return undefined; |
| } |
| }); |
| } |
| exports.detectGPUDevice = detectGPUDevice; |
| const canvasRenderWGSL = ` |
| @group(0) @binding(0) var my_sampler : sampler; |
| @group(0) @binding(1) var my_texture : texture_2d<f32>; |
| |
| struct VertexOutput { |
| @builtin(position) position : vec4<f32>, |
| @location(0) uv : vec2<f32>, |
| } |
| |
| @vertex |
| fn vertex_main(@builtin(vertex_index) vidx : u32) -> VertexOutput { |
| const pos = array( |
| vec2( 1.0, 1.0), |
| vec2( 1.0, -1.0), |
| vec2(-1.0, -1.0), |
| vec2( 1.0, 1.0), |
| vec2(-1.0, -1.0), |
| vec2(-1.0, 1.0), |
| ); |
| |
| const uv = array( |
| vec2(1.0, 0.0), |
| vec2(1.0, 1.0), |
| vec2(0.0, 1.0), |
| vec2(1.0, 0.0), |
| vec2(0.0, 1.0), |
| vec2(0.0, 0.0), |
| ); |
| |
| var output : VertexOutput; |
| output.position = vec4(pos[vidx], 0.0, 1.0); |
| output.uv = uv[vidx]; |
| return output; |
| } |
| |
| @fragment |
| fn fragment_main(@location(0) uv : vec2<f32>) -> @location(0) vec4<f32> { |
| return textureSample(my_texture, my_sampler, uv); |
| } |
| |
| @fragment |
| fn fragment_clear(@location(0) uv : vec2<f32>) -> @location(0) vec4<f32> { |
| return vec4(1.0, 1.0, 1.0, 1.0); |
| } |
| `; |
| class CanvaRenderManager { |
| constructor(device, canvas) { |
| this.device = device; |
| const ctx = canvas.getContext("webgpu", { |
| alpha: false, |
| antialias: false, |
| powerPreference: "high-performance", |
| }); |
| if (ctx == null) { |
| throw Error("Cannot bind WebGPU context"); |
| } |
| this.canvasContext = ctx; |
| this.canvasTextureFormat = navigator.gpu.getPreferredCanvasFormat(); |
| this.canvasContext.configure({ |
| device: this.device, |
| format: this.canvasTextureFormat, |
| alphaMode: "opaque", |
| }); |
| this.renderPipeline = device.createRenderPipeline({ |
| layout: "auto", |
| vertex: { |
| module: device.createShaderModule({ |
| code: canvasRenderWGSL, |
| }), |
| entryPoint: "vertex_main", |
| }, |
| fragment: { |
| module: device.createShaderModule({ |
| code: canvasRenderWGSL, |
| }), |
| entryPoint: "fragment_main", |
| targets: [{ |
| format: this.canvasTextureFormat, |
| }], |
| }, |
| primitive: { |
| topology: "triangle-list", |
| }, |
| }); |
| this.clearPipeline = device.createRenderPipeline({ |
| layout: "auto", |
| vertex: { |
| module: device.createShaderModule({ |
| code: canvasRenderWGSL, |
| }), |
| entryPoint: "vertex_main", |
| }, |
| fragment: { |
| module: device.createShaderModule({ |
| code: canvasRenderWGSL, |
| }), |
| entryPoint: "fragment_clear", |
| targets: [{ |
| format: this.canvasTextureFormat, |
| }], |
| }, |
| primitive: { |
| topology: "triangle-list", |
| }, |
| }); |
| this.renderSampler = device.createSampler({ |
| magFilter: "linear", |
| minFilter: "linear", |
| }); |
| |
| this.stagingTexture = device.createTexture({ |
| size: [canvas.height, canvas.width, 1], |
| format: "rgba8unorm", |
| usage: GPUTextureUsage.TEXTURE_BINDING | |
| GPUTextureUsage.COPY_DST | |
| GPUTextureUsage.RENDER_ATTACHMENT, |
| }); |
| } |
| clear() { |
| const commandEncoder = this.device.createCommandEncoder(); |
| const passEncoder = commandEncoder.beginRenderPass({ |
| colorAttachments: [ |
| { |
| view: this.canvasContext.getCurrentTexture().createView(), |
| clearValue: { r: 0.0, g: 0.0, b: 0.0, a: 1.0 }, |
| loadOp: "clear", |
| storeOp: "store", |
| }, |
| ], |
| }); |
| passEncoder.setPipeline(this.clearPipeline); |
| const renderBindingGroup = this.device.createBindGroup({ |
| layout: this.renderPipeline.getBindGroupLayout(0), |
| entries: [ |
| { binding: 0, resource: this.renderSampler }, |
| { binding: 1, resource: this.stagingTexture.createView() }, |
| ], |
| }); |
| passEncoder.setBindGroup(0, renderBindingGroup); |
| passEncoder.draw(6, 1, 0, 0); |
| passEncoder.end(); |
| this.device.queue.submit([commandEncoder.finish()]); |
| } |
| draw(buffer, height, width) { |
| |
| if (height != this.stagingTexture.height || width != this.stagingTexture.width) { |
| this.stagingTexture.destroy(); |
| this.stagingTexture = this.device.createTexture({ |
| size: [height, width, 1], |
| format: "rgba8unorm", |
| usage: GPUTextureUsage.TEXTURE_BINDING | |
| GPUTextureUsage.COPY_DST | |
| GPUTextureUsage.RENDER_ATTACHMENT, |
| }); |
| } |
| const commandEncoder = this.device.createCommandEncoder(); |
| commandEncoder.copyBufferToTexture({ |
| buffer: buffer, |
| offset: 0, |
| bytesPerRow: this.stagingTexture.width * 4 |
| }, { |
| texture: this.stagingTexture |
| }, { |
| width: this.stagingTexture.width, |
| height: this.stagingTexture.height |
| }); |
| const passEncoder = commandEncoder.beginRenderPass({ |
| colorAttachments: [ |
| { |
| view: this.canvasContext.getCurrentTexture().createView(), |
| clearValue: { r: 0.0, g: 0.0, b: 0.0, a: 1.0 }, |
| loadOp: "clear", |
| storeOp: "store", |
| }, |
| ], |
| }); |
| passEncoder.setPipeline(this.renderPipeline); |
| const renderBindingGroup = this.device.createBindGroup({ |
| layout: this.renderPipeline.getBindGroupLayout(0), |
| entries: [ |
| { binding: 0, resource: this.renderSampler }, |
| { binding: 1, resource: this.stagingTexture.createView() }, |
| ], |
| }); |
| passEncoder.setBindGroup(0, renderBindingGroup); |
| passEncoder.draw(6, 1, 0, 0); |
| passEncoder.end(); |
| this.device.queue.submit([commandEncoder.finish()]); |
| } |
| dispose() { |
| this.stagingTexture.destroy(); |
| } |
| } |
| |
| |
| |
| |
| class WebGPUContext { |
| constructor(memory, device) { |
| |
| this.bufferTable = [undefined]; |
| this.bufferTableFreeId = []; |
| this.podArgStagingBuffers = []; |
| this.canvasRenderManager = undefined; |
| |
| this.maxNumPodArgsStagingBuffers = 2; |
| |
| |
| |
| this.peakAllocatedBytes = 0; |
| |
| this.currAllocatedBytes = 0; |
| |
| this.allAllocatedBytes = 0; |
| |
| this.shaderSubmitCounter = 0; |
| |
| this.debugShaderSubmitLimit = -1; |
| |
| this.debugLogFinish = false; |
| this.memory = memory; |
| this.device = device; |
| } |
| |
| |
| |
| dispose() { |
| var _a, _b, _c; |
| (_a = this.canvasRenderManager) === null || _a === void 0 ? void 0 : _a.dispose(); |
| this.bufferTableFreeId = []; |
| while (this.bufferTable.length != 0) { |
| (_b = this.bufferTable.pop()) === null || _b === void 0 ? void 0 : _b.destroy(); |
| } |
| while (this.podArgStagingBuffers.length != 0) { |
| (_c = this.podArgStagingBuffers.pop()) === null || _c === void 0 ? void 0 : _c.destroy(); |
| } |
| this.device.destroy(); |
| } |
| |
| |
| |
| sync() { |
| return __awaiter(this, void 0, void 0, function* () { |
| yield this.device.queue.onSubmittedWorkDone(); |
| }); |
| } |
| |
| |
| |
| runtimeStatsText() { |
| let info = "peak-memory=" + Math.ceil(this.peakAllocatedBytes / (1 << 20)) + " MB"; |
| info += ", all-memory=" + Math.ceil(this.allAllocatedBytes / (1 << 20)) + " MB"; |
| info += ", shader-submissions=" + this.shaderSubmitCounter; |
| return info; |
| } |
| |
| |
| |
| |
| |
| |
| drawImageFromBuffer(ptr, height, width) { |
| if (this.canvasRenderManager == undefined) { |
| throw Error("Do not have a canvas context, call bindCanvas first"); |
| } |
| this.canvasRenderManager.draw(this.gpuBufferFromPtr(ptr), height, width); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| copyRawBytesToBuffer(rawBytes, toPtr, toOffset, nbytes) { |
| |
| this.device.queue.writeBuffer(this.gpuBufferFromPtr(toPtr), toOffset, rawBytes, 0, nbytes); |
| } |
| |
| |
| |
| clearCanvas() { |
| var _a; |
| (_a = this.canvasRenderManager) === null || _a === void 0 ? void 0 : _a.clear(); |
| } |
| |
| |
| |
| |
| bindCanvas(canvas) { |
| this.canvasRenderManager = new CanvaRenderManager(this.device, canvas); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| createShader(finfo, code) { |
| return this.createShadeInternal(finfo, code, false); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| createShaderAsync(finfo, code) { |
| return __awaiter(this, void 0, void 0, function* () { |
| return yield this.createShadeInternal(finfo, code, true); |
| }); |
| } |
| |
| |
| |
| |
| |
| getPodArgsBuffer(nbytes) { |
| let buffer = undefined; |
| if (this.podArgStagingBuffers.length >= this.maxNumPodArgsStagingBuffers) { |
| buffer = this.podArgStagingBuffers.shift(); |
| } |
| |
| let allocSize = 16; |
| if (buffer !== undefined) { |
| allocSize = buffer.size; |
| if (buffer.size < nbytes) { |
| buffer.destroy(); |
| buffer = undefined; |
| } |
| } |
| while (allocSize < nbytes) { |
| allocSize *= 2; |
| } |
| if (buffer == undefined) { |
| |
| buffer = this.device.createBuffer({ |
| size: allocSize, |
| usage: GPUBufferUsage.UNIFORM | GPUBufferUsage.COPY_DST, |
| }); |
| } |
| support.assert(nbytes <= buffer.size); |
| return buffer; |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| createShadeInternal(finfo, code, asyncMode) { |
| const dispatchToDim = []; |
| let paramWriteAccess = []; |
| for (let i = 0; i < finfo.launch_param_tags.length; ++i) { |
| const tag = finfo.launch_param_tags[i]; |
| if (tag.startsWith("blockIdx.")) { |
| const target = tag.charCodeAt(tag.length - 1) - ("x".charCodeAt(0)); |
| support.assert(target >= 0 && target < 3); |
| dispatchToDim.push(target); |
| } |
| else if (tag.startsWith("threadIdx.")) { |
| const target = tag.charCodeAt(tag.length - 1) - ("x".charCodeAt(0)); |
| support.assert(target >= 0 && target < 3); |
| dispatchToDim.push(target + 3); |
| } |
| else if (tag.startsWith("paramWriteAccess:")) { |
| paramWriteAccess = JSON.parse(tag.substring(17)); |
| } |
| else { |
| throw new Error("Cannot handle thread_axis " + tag); |
| } |
| } |
| const layoutEntries = []; |
| const bufferArgIndices = []; |
| const podArgIndices = []; |
| for (let i = 0; i < finfo.arg_types.length; ++i) { |
| const dtype = finfo.arg_types[i]; |
| if (dtype == "handle") { |
| layoutEntries.push({ |
| binding: bufferArgIndices.length, |
| visibility: GPUShaderStage.COMPUTE, |
| buffer: { |
| type: paramWriteAccess[bufferArgIndices.length] ? "storage" : "read-only-storage" |
| } |
| }); |
| bufferArgIndices.push(i); |
| } |
| else if (dtype.startsWith("int") || dtype.startsWith("uint") || dtype.startsWith("float")) { |
| podArgIndices.push(i); |
| } |
| else { |
| throw new Error("Cannot handle argument type " + dtype + " in WebGPU shader"); |
| } |
| } |
| support.assert(paramWriteAccess.length == bufferArgIndices.length); |
| |
| layoutEntries.push({ |
| binding: bufferArgIndices.length, |
| visibility: GPUShaderStage.COMPUTE, |
| buffer: { |
| type: "uniform" |
| } |
| }); |
| const bindGroupLayout = this.device.createBindGroupLayout({ |
| entries: layoutEntries |
| }); |
| const pipelineLayout = this.device.createPipelineLayout({ |
| bindGroupLayouts: [bindGroupLayout] |
| }); |
| |
| const createShaderFunc = (pipeline) => { |
| const submitShader = (...args) => { |
| if (this.debugShaderSubmitLimit != -1 && |
| this.shaderSubmitCounter >= this.debugShaderSubmitLimit) { |
| this.shaderSubmitCounter += 1; |
| return; |
| } |
| const commandEncoder = this.device.createCommandEncoder(); |
| const compute = commandEncoder.beginComputePass(); |
| compute.setPipeline(pipeline); |
| const bindGroupEntries = []; |
| const numBufferOrPodArgs = bufferArgIndices.length + podArgIndices.length; |
| support.assert(args.length == numBufferOrPodArgs + dispatchToDim.length); |
| const workDim = [1, 1, 1, 1, 1, 1]; |
| for (let i = 0; i < dispatchToDim.length; ++i) { |
| workDim[dispatchToDim[i]] = args[numBufferOrPodArgs + i]; |
| } |
| |
| if (workDim[2] != 1) { |
| throw Error("WebGPU: blockIdx.z is reserved for internal use"); |
| } |
| const packDimX = workDim[0]; |
| |
| if (workDim[0] >= (1 << 16)) { |
| let wl_x = workDim[0]; |
| let wl_z = workDim[2]; |
| while (wl_x >= (1 << 16)) { |
| if (wl_x % 2 == 0) { |
| wl_x = wl_x / 2; |
| } |
| else { |
| |
| wl_x = (wl_x + 1) / 2; |
| } |
| wl_z *= 2; |
| } |
| workDim[0] = wl_x; |
| workDim[2] = wl_z; |
| support.assert(wl_x * wl_z >= packDimX); |
| } |
| for (let i = 0; i < bufferArgIndices.length; ++i) { |
| bindGroupEntries.push({ |
| binding: i, |
| resource: { |
| buffer: this.gpuBufferFromPtr(args[bufferArgIndices[i]]) |
| } |
| }); |
| } |
| |
| const sizeOfI32 = 4; |
| const podArgBuffer = this.getPodArgsBuffer((podArgIndices.length + 1) * sizeOfI32); |
| const i32View = new Int32Array(podArgIndices.length + 1); |
| const u32View = new Uint32Array(i32View.buffer); |
| const f32View = new Float32Array(i32View.buffer); |
| for (let i = 0; i < podArgIndices.length; ++i) { |
| const value = args[podArgIndices[i]]; |
| const dtype = finfo.arg_types[podArgIndices[i]]; |
| if (dtype.startsWith("int")) { |
| i32View[i] = value; |
| } |
| else if (dtype.startsWith("uint")) { |
| u32View[i] = value; |
| } |
| else if (dtype.startsWith("float")) { |
| f32View[i] = value; |
| } |
| else { |
| throw Error("Unknown pod dtype " + dtype); |
| } |
| } |
| |
| u32View[podArgIndices.length] = packDimX; |
| this.device.queue.writeBuffer(podArgBuffer, 0, i32View.buffer); |
| bindGroupEntries.push({ |
| binding: bufferArgIndices.length, |
| resource: { |
| buffer: podArgBuffer, |
| size: i32View.buffer.byteLength |
| } |
| }); |
| compute.setBindGroup(0, this.device.createBindGroup({ |
| layout: bindGroupLayout, |
| entries: bindGroupEntries |
| })); |
| compute.dispatchWorkgroups(workDim[0], workDim[1], workDim[2]); |
| compute.end(); |
| const command = commandEncoder.finish(); |
| this.device.queue.submit([command]); |
| if (this.debugLogFinish) { |
| const currCounter = this.shaderSubmitCounter; |
| this.device.queue.onSubmittedWorkDone().then(() => { |
| console.log("[" + currCounter + "][Debug] finish shader" + finfo.name); |
| }); |
| } |
| this.shaderSubmitCounter += 1; |
| }; |
| return submitShader; |
| }; |
| const shaderModule = this.device.createShaderModule({ |
| code: code, |
| hints: { |
| main: { |
| layout: pipelineLayout |
| } |
| } |
| }); |
| if (asyncMode) { |
| return this.device.createComputePipelineAsync({ |
| layout: pipelineLayout, |
| compute: { |
| module: shaderModule, |
| entryPoint: finfo.name |
| } |
| }).then((pipeline) => { |
| return createShaderFunc(pipeline); |
| }); |
| } |
| else { |
| const pipeline = this.device.createComputePipeline({ |
| layout: pipelineLayout, |
| compute: { |
| module: shaderModule, |
| entryPoint: finfo.name |
| } |
| }); |
| return createShaderFunc(pipeline); |
| } |
| } |
| |
| |
| |
| |
| |
| getDeviceAPI(name) { |
| if (name == "deviceAllocDataSpace") { |
| return (nbytes) => { |
| return this.deviceAllocDataSpace(nbytes); |
| }; |
| } |
| else if (name == "deviceFreeDataSpace") { |
| return (ptr) => { |
| return this.deviceFreeDataSpace(ptr); |
| }; |
| } |
| else if (name == "deviceCopyToGPU") { |
| return (from, to, toOffset, nbytes) => { |
| this.deviceCopyToGPU(from, to, toOffset, nbytes); |
| }; |
| } |
| else if (name == "deviceCopyFromGPU") { |
| return (from, fromOffset, to, nbytes) => { |
| this.deviceCopyFromGPU(from, fromOffset, to, nbytes); |
| }; |
| } |
| else if (name == "deviceCopyWithinGPU") { |
| return (from, fromOffset, to, toOffset, nbytes) => { |
| this.deviceCopyWithinGPU(from, fromOffset, to, toOffset, nbytes); |
| }; |
| } |
| else { |
| throw new Error("Unknown DeviceAPI function " + name); |
| } |
| } |
| |
| deviceAllocDataSpace(nbytes) { |
| |
| if (nbytes == 0) { |
| nbytes = 1; |
| } |
| const buffer = this.device.createBuffer({ |
| size: nbytes, |
| usage: GPUBufferUsage.STORAGE | GPUBufferUsage.COPY_SRC | GPUBufferUsage.COPY_DST, |
| }); |
| this.currAllocatedBytes += nbytes; |
| this.allAllocatedBytes += nbytes; |
| if (this.currAllocatedBytes > this.peakAllocatedBytes) { |
| this.peakAllocatedBytes = this.currAllocatedBytes; |
| } |
| const ptr = this.attachToBufferTable(buffer); |
| return ptr; |
| } |
| deviceFreeDataSpace(ptr) { |
| const idx = ptr; |
| const buffer = this.bufferTable[idx]; |
| this.bufferTable[idx] = undefined; |
| support.assert(buffer !== undefined); |
| this.bufferTableFreeId.push(idx); |
| this.currAllocatedBytes -= buffer.size; |
| buffer.destroy(); |
| } |
| deviceCopyToGPU(from, to, toOffset, nbytes) { |
| |
| const rawBytes = this.memory.loadRawBytes(from, nbytes); |
| this.device.queue.writeBuffer(this.gpuBufferFromPtr(to), toOffset, rawBytes, 0, nbytes); |
| } |
| deviceCopyFromGPU(from, fromOffset, to, nbytes) { |
| |
| const gpuTemp = this.device.createBuffer({ |
| size: nbytes, |
| usage: GPUBufferUsage.MAP_READ | GPUBufferUsage.COPY_DST, |
| }); |
| const copyEncoder = this.device.createCommandEncoder(); |
| copyEncoder.copyBufferToBuffer(this.gpuBufferFromPtr(from), fromOffset, gpuTemp, 0, nbytes); |
| const copyCommands = copyEncoder.finish(); |
| this.device.queue.submit([copyCommands]); |
| gpuTemp.mapAsync(GPUMapMode.READ).then(() => { |
| const data = gpuTemp.getMappedRange(); |
| this.memory.storeRawBytes(to, new Uint8Array(data)); |
| gpuTemp.destroy(); |
| }); |
| } |
| deviceCopyWithinGPU(from, fromOffset, to, toOffset, nbytes) { |
| const copyEncoder = this.device.createCommandEncoder(); |
| copyEncoder.copyBufferToBuffer(this.gpuBufferFromPtr(from), fromOffset, this.gpuBufferFromPtr(to), toOffset, nbytes); |
| const copyCommands = copyEncoder.finish(); |
| this.device.queue.submit([copyCommands]); |
| } |
| gpuBufferFromPtr(ptr) { |
| const buffer = this.bufferTable[ptr]; |
| support.assert(buffer !== undefined); |
| return buffer; |
| } |
| attachToBufferTable(buffer) { |
| if (this.bufferTableFreeId.length != 0) { |
| const idx = this.bufferTableFreeId.pop(); |
| this.bufferTable[idx] = buffer; |
| return idx; |
| } |
| else { |
| const idx = this.bufferTable.length; |
| this.bufferTable.push(buffer); |
| return idx; |
| } |
| } |
| } |
| exports.WebGPUContext = WebGPUContext; |
|
|
| }); |
|
|
| unwrapExports(webgpu); |
| webgpu.WebGPUContext; |
| webgpu.detectGPUDevice; |
|
|
| var compact = createCommonjsModule(function (module, exports) { |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| Object.defineProperty(exports, "__esModule", { value: true }); |
| exports.createWebSocket = exports.getPerformance = void 0; |
| |
| |
| |
| function getPerformance() { |
| if (typeof performance == "undefined") { |
| |
| const performanceNode = perf_hooks__default["default"]; |
| return performanceNode.performance; |
| } |
| else { |
| return performance; |
| } |
| } |
| exports.getPerformance = getPerformance; |
| |
| |
| |
| |
| function createWebSocket(url) { |
| if (typeof WebSocket == "undefined") { |
| |
| const WebSocket = ws__default["default"]; |
| return new WebSocket(url); |
| } |
| else { |
| return new WebSocket(url); |
| } |
| } |
| exports.createWebSocket = createWebSocket; |
|
|
| }); |
|
|
| unwrapExports(compact); |
| compact.createWebSocket; |
| compact.getPerformance; |
|
|
| var runtime = createCommonjsModule(function (module, exports) { |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| var __awaiter = (commonjsGlobal && commonjsGlobal.__awaiter) || function (thisArg, _arguments, P, generator) { |
| function adopt(value) { return value instanceof P ? value : new P(function (resolve) { resolve(value); }); } |
| return new (P || (P = Promise))(function (resolve, reject) { |
| function fulfilled(value) { try { step(generator.next(value)); } catch (e) { reject(e); } } |
| function rejected(value) { try { step(generator["throw"](value)); } catch (e) { reject(e); } } |
| function step(result) { result.done ? resolve(result.value) : adopt(result.value).then(fulfilled, rejected); } |
| step((generator = generator.apply(thisArg, _arguments || [])).next()); |
| }); |
| }; |
| Object.defineProperty(exports, "__esModule", { value: true }); |
| exports.instantiate = exports.Instance = exports.VirtualMachine = exports.TVMArray = exports.TVMObject = exports.Module = exports.NDArray = exports.DLDataType = exports.DLDevice = exports.Scalar = void 0; |
|
|
|
|
|
|
|
|
|
|
| |
| |
| |
| |
| class FFILibrary { |
| constructor(wasmInstance, imports) { |
| this.recycledCallStacks = []; |
| this.wasmInstance = wasmInstance; |
| this.memory = new memory.Memory(this.detectWasmMemory(this.wasmInstance, imports)); |
| support.assert(this.wasmInstance.exports !== undefined, "Expect the library module contains exports"); |
| this.exports = this.wasmInstance.exports; |
| this.wasm32 = this.memory.wasm32; |
| this.validateInstance(); |
| } |
| dispose() { |
| var _a; |
| while (this.recycledCallStacks.length != 0) { |
| this.recycledCallStacks.pop().dispose(); |
| } |
| (_a = this.webGPUContext) === null || _a === void 0 ? void 0 : _a.dispose(); |
| } |
| sizeofPtr() { |
| return this.memory.sizeofPtr(); |
| } |
| checkCall(code) { |
| if (code != 0) { |
| const msgPtr = this.exports |
| .TVMGetLastError(); |
| throw new Error("TVMError: " + this.memory.loadCString(msgPtr)); |
| } |
| } |
| getOrAllocCallStack() { |
| if (this.recycledCallStacks.length != 0) { |
| return this.recycledCallStacks.pop(); |
| } |
| return new memory.CachedCallStack(this.memory, this.exports.TVMWasmAllocSpace, this.exports.TVMWasmFreeSpace); |
| } |
| recycleCallStack(callstack) { |
| callstack.reset(); |
| this.recycledCallStacks.push(callstack); |
| } |
| validateInstance() { |
| this.checkExports(["TVMWasmAllocSpace", "TVMWasmFreeSpace", "TVMFuncFree"]); |
| } |
| checkExports(funcNames) { |
| const missList = []; |
| for (const name of funcNames) { |
| const f = this.exports[name]; |
| if (!(f instanceof Function)) { |
| missList.push(name); |
| } |
| } |
| if (missList.length != 0) { |
| throw new Error("Cannot find " + missList + " in exports"); |
| } |
| } |
| detectWasmMemory(instance, imports) { |
| if (instance.exports.memory instanceof WebAssembly.Memory) { |
| return instance.exports.memory; |
| } |
| if (imports.env && imports.env.memory instanceof WebAssembly.Memory) { |
| return imports.env.memory; |
| } |
| throw new Error("Cannt detect wasm memory from imports " + |
| imports + |
| " or exports" + |
| instance.exports); |
| } |
| } |
| |
| |
| |
| |
| class RuntimeContext { |
| constructor(getGlobalFunc) { |
| this.autoDisposeScope = []; |
| this.arrayGetItem = getGlobalFunc("runtime.ArrayGetItem"); |
| this.arrayGetSize = getGlobalFunc("runtime.ArraySize"); |
| this.arrayMake = getGlobalFunc("runtime.Array"); |
| this.getSysLib = getGlobalFunc("runtime.SystemLib"); |
| this.arrayCacheGet = getGlobalFunc("vm.builtin.ndarray_cache.get"); |
| this.arrayCacheRemove = getGlobalFunc("vm.builtin.ndarray_cache.remove"); |
| this.arrayCacheUpdate = getGlobalFunc("vm.builtin.ndarray_cache.update"); |
| this.arrayCacheClear = getGlobalFunc("vm.builtin.ndarray_cache.clear"); |
| this.arrayDecodeStorage = getGlobalFunc("tvmjs.array.decode_storage"); |
| this.paramModuleFromCache = getGlobalFunc("vm.builtin.param_module_from_cache"); |
| this.makeShapeTuple = getGlobalFunc("runtime.ShapeTuple"); |
| this.ndarrayCreateView = getGlobalFunc("runtime.TVMArrayCreateView"); |
| this.sampleTopPFromLogits = getGlobalFunc("vm.builtin.sample_top_p_from_logits"); |
| } |
| dispose() { |
| |
| this.arrayCacheClear.dispose(); |
| this.arrayGetItem.dispose(); |
| this.arrayGetSize.dispose(); |
| this.arrayMake.dispose(); |
| this.arrayCacheGet.dispose(); |
| this.arrayCacheRemove.dispose(); |
| this.arrayCacheUpdate.dispose(); |
| this.arrayCacheClear.dispose(); |
| this.arrayDecodeStorage.dispose(); |
| this.paramModuleFromCache.dispose(); |
| this.makeShapeTuple.dispose(); |
| this.ndarrayCreateView.dispose(); |
| this.sampleTopPFromLogits.dispose(); |
| } |
| beginScope() { |
| this.autoDisposeScope.push([]); |
| } |
| endScope() { |
| if (this.autoDisposeScope.length == 0) { |
| throw Error("tvm.endScope called when the stack is empty."); |
| } |
| |
| const currScope = this.autoDisposeScope.pop(); |
| for (let i = 0; i < currScope.length; ++i) { |
| const val = currScope[i]; |
| if (val !== undefined) { |
| val.dispose(); |
| } |
| } |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| attachToCurrentScope(obj) { |
| if (this.autoDisposeScope.length == 0) { |
| throw Error("Must call beginScope to use functions that returns TVM objects"); |
| } |
| const currScope = this.autoDisposeScope[this.autoDisposeScope.length - 1]; |
| currScope.push(obj); |
| return obj; |
| } |
| moveToParentScope(obj) { |
| this.detachFromCurrentScope(obj); |
| if (this.autoDisposeScope.length < 2) { |
| throw Error("moveToParentScope: Parent scope do not exist"); |
| } |
| const parentScope = this.autoDisposeScope[this.autoDisposeScope.length - 2]; |
| parentScope.push(obj); |
| return obj; |
| } |
| detachFromCurrentScope(obj) { |
| const currScope = this.autoDisposeScope[this.autoDisposeScope.length - 1]; |
| let occurance = 0; |
| for (let i = 0; i < currScope.length; ++i) { |
| if (currScope[i] === obj) { |
| occurance += 1; |
| currScope[i] = undefined; |
| } |
| } |
| if (occurance == 0) { |
| throw Error("Cannot find obj in the current auto conversion pool"); |
| } |
| if (occurance > 1) { |
| throw Error("Value attached to scope multiple times"); |
| } |
| return obj; |
| } |
| } |
| |
| |
| |
| |
| class Scalar { |
| constructor(value, dtype) { |
| this.value = value; |
| this.dtype = dtype; |
| } |
| } |
| exports.Scalar = Scalar; |
| |
| |
| |
| class PackedFuncCell { |
| constructor(handle, lib) { |
| this.handle = handle; |
| this.lib = lib; |
| } |
| dispose() { |
| if (this.handle != 0) { |
| this.lib.checkCall(this.lib.exports.TVMFuncFree(this.handle)); |
| this.handle = 0; |
| } |
| } |
| getHandle(requireNotNull = true) { |
| if (requireNotNull && this.handle == 0) { |
| throw Error("PackedFunc has already been disposed"); |
| } |
| return this.handle; |
| } |
| } |
| const DeviceEnumToStr = { |
| 1: "cpu", |
| 2: "cuda", |
| 4: "opencl", |
| 8: "metal", |
| 15: "webgpu" |
| }; |
| const DeviceStrToEnum = { |
| cpu: 1, |
| cuda: 2, |
| cl: 4, |
| opencl: 4, |
| vulkan: 7, |
| metal: 8, |
| webgpu: 15 |
| }; |
| |
| |
| |
| class DLDevice { |
| constructor(deviceType, deviceId, lib) { |
| const tp = typeof deviceType; |
| if (tp == "string") { |
| this.deviceType = DeviceStrToEnum[deviceType]; |
| if (this.deviceType == undefined) { |
| throw new Error("Cannot recogonize deviceType " + deviceType); |
| } |
| } |
| else if (tp == "number") { |
| this.deviceType = deviceType; |
| } |
| else { |
| throw new Error("Cannot take type " + tp + " as deviceType"); |
| } |
| this.deviceId = deviceId; |
| this.lib = lib; |
| } |
| |
| |
| |
| sync() { |
| return __awaiter(this, void 0, void 0, function* () { |
| if (this.deviceType == DeviceStrToEnum.webgpu) { |
| support.assert(this.lib.webGPUContext !== undefined); |
| yield this.lib.webGPUContext.sync(); |
| } |
| }); |
| } |
| toString() { |
| return (DeviceEnumToStr[this.deviceType] + "(" + this.deviceId.toString() + ")"); |
| } |
| } |
| exports.DLDevice = DLDevice; |
| const DLDataTypeCodeToStr = { |
| 0: "int", |
| 1: "uint", |
| 2: "float", |
| 3: "handle", |
| }; |
| |
| |
| |
| class DLDataType { |
| constructor(code, bits, lanes) { |
| this.code = code; |
| this.bits = bits; |
| this.lanes = lanes; |
| } |
| toString() { |
| const ret = DLDataTypeCodeToStr[this.code] + this.bits.toString(); |
| if (this.lanes != 1) { |
| return ret + "x" + this.lanes.toString(); |
| } |
| else { |
| return ret; |
| } |
| } |
| numStorageBytes() { |
| return (this.bits * this.lanes + 7) >> 3; |
| } |
| } |
| exports.DLDataType = DLDataType; |
| |
| |
| |
| class NDArray { |
| constructor(handle, isView, lib, ctx) { |
| this.handle = handle; |
| this.isView = isView; |
| this.lib = lib; |
| this.ctx = ctx; |
| if (this.isView) { |
| this.dltensor = handle; |
| } |
| else { |
| this.dltensor = this.getDLTensorFromArrayHandle(this.handle); |
| } |
| |
| const arrayOffsetData = 0; |
| const arrayOffsetContext = arrayOffsetData + this.lib.sizeofPtr(); |
| const arrayOffsetDevType = arrayOffsetContext; |
| const arrayOffsetDevId = arrayOffsetContext + 4 ; |
| const arrayOffsetNdim = arrayOffsetContext + 8 ; |
| const arrayOffsetDtype = arrayOffsetNdim + 4 ; |
| const arrayOffsetDtypeCode = arrayOffsetDtype; |
| const arrayOffsetDtypeBits = arrayOffsetDtype + 1 ; |
| const arrayOffsetDtypeLanes = arrayOffsetDtypeBits + 1 ; |
| const arrayOffsetShape = arrayOffsetDtype + 4 ; |
| const arrayOffsetStrides = arrayOffsetShape + this.lib.sizeofPtr(); |
| const arrayOffsetByteOffset = arrayOffsetStrides + this.lib.sizeofPtr(); |
| |
| this.dataPtr = lib.memory.loadPointer(this.dltensor); |
| |
| this.ndim = lib.memory.loadI32(this.dltensor + arrayOffsetNdim); |
| |
| const cshapePtr = lib.memory.loadPointer(this.dltensor + arrayOffsetShape); |
| this.shape = []; |
| for (let i = 0; i < this.ndim; ++i) { |
| this.shape.push(lib.memory.loadI64(cshapePtr + i * 8 )); |
| } |
| |
| const code = lib.memory.loadU8(this.dltensor + arrayOffsetDtypeCode); |
| const bits = lib.memory.loadU8(this.dltensor + arrayOffsetDtypeBits); |
| const lanes = lib.memory.loadU16(this.dltensor + arrayOffsetDtypeLanes); |
| this.dlDataType = new DLDataType(code, bits, lanes); |
| this.dtype = this.dlDataType.toString(); |
| |
| const deviceType = lib.memory.loadI32(this.dltensor + arrayOffsetDevType); |
| const deviceId = lib.memory.loadI32(this.dltensor + arrayOffsetDevId); |
| this.device = new DLDevice(deviceType, deviceId, lib); |
| |
| this.byteOffset = lib.memory.loadI64(this.dltensor + arrayOffsetByteOffset); |
| } |
| |
| |
| |
| |
| |
| view(shape) { |
| const shapeArray = shape.map((value) => new Scalar(value, "int")); |
| return this.ctx.ndarrayCreateView(this, this.ctx.makeShapeTuple(...shapeArray)); |
| } |
| |
| |
| |
| |
| |
| |
| getHandle(requireNotNull = true) { |
| if (requireNotNull && this.handle == 0) { |
| throw Error("NDArray has already been disposed"); |
| } |
| return this.handle; |
| } |
| |
| |
| |
| |
| |
| getDataPtr() { |
| if (this.handle == 0) { |
| throw Error("NDArray has already been disposed"); |
| } |
| return this.dataPtr; |
| } |
| dispose() { |
| if (this.handle != 0 && !this.isView) { |
| this.lib.checkCall(this.lib.exports.TVMArrayFree(this.handle)); |
| this.handle = 0; |
| } |
| } |
| |
| |
| |
| |
| |
| |
| |
| copyFrom(data) { |
| if (data instanceof NDArray) { |
| this.lib.checkCall(this.lib.exports.TVMArrayCopyFromTo(data.getHandle(), this.getHandle(), 0)); |
| return this; |
| } |
| else { |
| const size = this.shape.reduce((a, b) => { |
| return a * b; |
| }, 1); |
| if (data.length != size) { |
| throw new Error("data size and shape mismatch data.length" + |
| data.length + |
| " vs " + |
| size); |
| } |
| let buffer; |
| if (this.dtype == "float32") { |
| buffer = Float32Array.from(data).buffer; |
| } |
| else if (this.dtype == "float64") { |
| buffer = Float64Array.from(data).buffer; |
| } |
| else if (this.dtype == "int32") { |
| buffer = Int32Array.from(data).buffer; |
| } |
| else if (this.dtype == "int8") { |
| buffer = Int8Array.from(data).buffer; |
| } |
| else if (this.dtype == "uint8") { |
| buffer = Uint8Array.from(data).buffer; |
| } |
| else { |
| throw new Error("Unsupported data type " + this.dtype); |
| } |
| return this.copyFromRawBytes(new Uint8Array(buffer)); |
| } |
| } |
| |
| |
| |
| |
| |
| copyFromRawBytes(data) { |
| var _a; |
| |
| if (this.device.deviceType == DeviceStrToEnum.webgpu) { |
| (_a = this.lib.webGPUContext) === null || _a === void 0 ? void 0 : _a.copyRawBytesToBuffer(data, this.getDataPtr(), 0, data.length); |
| return this; |
| } |
| |
| const size = this.shape.reduce((a, b) => { |
| return a * b; |
| }, 1); |
| const nbytes = this.dlDataType.numStorageBytes() * size; |
| if (nbytes != data.length) { |
| throw new Error("Expect the data's length equals nbytes=" + nbytes); |
| } |
| const stack = this.lib.getOrAllocCallStack(); |
| const tempOffset = stack.allocRawBytes(nbytes); |
| const tempPtr = stack.ptrFromOffset(tempOffset); |
| this.lib.memory.storeRawBytes(tempPtr, data); |
| this.lib.checkCall(this.lib.exports.TVMArrayCopyFromBytes(this.getHandle(), tempPtr, nbytes)); |
| this.lib.recycleCallStack(stack); |
| return this; |
| } |
| |
| |
| |
| |
| toRawBytes() { |
| if (this.device.deviceType != DeviceStrToEnum.cpu) { |
| throw new Error("Can only sync copy CPU array, use cpu_arr.copyfrom(gpu_arr) then sync instead."); |
| } |
| const size = this.shape.reduce((a, b) => { |
| return a * b; |
| }, 1); |
| const nbytes = this.dlDataType.numStorageBytes() * size; |
| const stack = this.lib.getOrAllocCallStack(); |
| const tempOffset = stack.allocRawBytes(nbytes); |
| const tempPtr = stack.ptrFromOffset(tempOffset); |
| this.lib.checkCall(this.lib.exports.TVMArrayCopyToBytes(this.getHandle(), tempPtr, nbytes)); |
| const ret = this.lib.memory.loadRawBytes(tempPtr, nbytes); |
| this.lib.recycleCallStack(stack); |
| return ret; |
| } |
| |
| |
| |
| |
| |
| toArray() { |
| const stype = this.dtype; |
| if (stype == "float32") { |
| return new Float32Array(this.toRawBytes().buffer); |
| } |
| else if (stype == "float64") { |
| return new Float64Array(this.toRawBytes().buffer); |
| } |
| else if (stype == "int32") { |
| return new Int32Array(this.toRawBytes().buffer); |
| } |
| else if (stype == "int8") { |
| return new Int8Array(this.toRawBytes().buffer); |
| } |
| else if (stype == "uint8") { |
| return new Uint8Array(this.toRawBytes().buffer); |
| } |
| else { |
| throw new Error("Unsupported data type " + this.dtype); |
| } |
| } |
| getDLTensorFromArrayHandle(handle) { |
| |
| |
| return handle; |
| } |
| } |
| exports.NDArray = NDArray; |
| |
| |
| |
| class Module { |
| constructor(handle, lib, makePackedFunc) { |
| this.handle = handle; |
| this.lib = lib; |
| this.makePackedFunc = makePackedFunc; |
| } |
| dispose() { |
| if (this.handle != 0) { |
| this.lib.checkCall(this.lib.exports.TVMModFree(this.handle)); |
| this.handle = 0; |
| } |
| } |
| |
| |
| |
| |
| |
| |
| getHandle(requireNotNull = true) { |
| if (requireNotNull && this.handle == 0) { |
| throw Error("Module has already been disposed"); |
| } |
| return this.handle; |
| } |
| |
| |
| |
| |
| |
| |
| getFunction(name, queryImports = true) { |
| if (this.handle == 0) { |
| throw Error("Module has already been disposed"); |
| } |
| const stack = this.lib.getOrAllocCallStack(); |
| const nameOffset = stack.allocRawBytes(name.length + 1); |
| stack.storeRawBytes(nameOffset, support.StringToUint8Array(name)); |
| const outOffset = stack.allocPtrArray(1); |
| const outPtr = stack.ptrFromOffset(outOffset); |
| stack.commitToWasmMemory(outOffset); |
| this.lib.checkCall(this.lib.exports.TVMModGetFunction(this.getHandle(), stack.ptrFromOffset(nameOffset), queryImports ? 1 : 0, outPtr)); |
| const handle = this.lib.memory.loadPointer(outPtr); |
| this.lib.recycleCallStack(stack); |
| if (handle == 0) { |
| throw Error("Cannot find function " + name); |
| } |
| const ret = this.makePackedFunc(handle); |
| return ret; |
| } |
| |
| |
| |
| |
| importModule(mod) { |
| this.lib.checkCall(this.lib.exports.TVMModImport(this.getHandle(), mod.getHandle())); |
| } |
| } |
| exports.Module = Module; |
| |
| |
| |
| class TVMObject { |
| constructor(handle, lib, ctx) { |
| this.handle = handle; |
| this.lib = lib; |
| this.ctx = ctx; |
| } |
| dispose() { |
| if (this.handle != 0) { |
| this.lib.checkCall(this.lib.exports.TVMObjectFree(this.handle)); |
| this.handle = 0; |
| } |
| } |
| |
| |
| |
| |
| |
| |
| getHandle(requireNotNull = true) { |
| if (requireNotNull && this.handle == 0) { |
| throw Error("Module has already been disposed"); |
| } |
| return this.handle; |
| } |
| |
| typeIndex() { |
| if (this.handle == 0) { |
| throw Error("The current Object has already been disposed"); |
| } |
| const stack = this.lib.getOrAllocCallStack(); |
| const outOffset = stack.allocPtrArray(1); |
| const outPtr = stack.ptrFromOffset(outOffset); |
| this.lib.checkCall(this.lib.exports.TVMObjectGetTypeIndex(this.getHandle(), outPtr)); |
| const result = this.lib.memory.loadU32(outPtr); |
| this.lib.recycleCallStack(stack); |
| return result; |
| } |
| |
| typeKey() { |
| const type_index = this.typeIndex(); |
| const stack = this.lib.getOrAllocCallStack(); |
| const outOffset = stack.allocPtrArray(1); |
| const outPtr = stack.ptrFromOffset(outOffset); |
| this.lib.checkCall(this.lib.exports.TVMObjectTypeIndex2Key(type_index, outPtr)); |
| const result = this.lib.memory.loadCString(this.lib.memory.loadPointer(outPtr)); |
| this.lib.recycleCallStack(stack); |
| return result; |
| } |
| } |
| exports.TVMObject = TVMObject; |
| |
| class TVMArray extends TVMObject { |
| constructor(handle, lib, ctx) { |
| super(handle, lib, ctx); |
| } |
| |
| |
| |
| size() { |
| return this.ctx.arrayGetSize(this); |
| } |
| |
| |
| |
| |
| |
| get(index) { |
| return this.ctx.arrayGetItem(this, new Scalar(index, "int32")); |
| } |
| } |
| exports.TVMArray = TVMArray; |
| |
| |
| |
| |
| |
| |
| |
| class VirtualMachine { |
| |
| |
| |
| |
| |
| constructor(mod, device) { |
| this.mod = mod; |
| this.mod.getFunction("vm_initialization")(new Scalar(device.deviceType, "int"), new Scalar(device.deviceId, "int"), new Scalar(2 , "int"), |
| |
| new Scalar(DeviceStrToEnum.cpu, "int"), new Scalar(0, "int"), new Scalar(2 , "int")); |
| } |
| dispose() { |
| this.mod.dispose(); |
| } |
| |
| |
| |
| |
| |
| getFunction(name) { |
| return this.mod.getFunction(name); |
| } |
| |
| |
| |
| getInternalModule() { |
| return this.mod; |
| } |
| } |
| exports.VirtualMachine = VirtualMachine; |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| class Instance { |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| constructor(wasmModule, importObject = {}, wasmInstance, env) { |
| this.cacheMetadata = {}; |
| this.initProgressCallback = []; |
| if (wasmInstance instanceof WebAssembly.Instance) { |
| support.assert(env instanceof environment.Environment, "env must be provided when passing in instance"); |
| } |
| else { |
| support.assert(env === undefined); |
| env = new environment.Environment(importObject); |
| wasmInstance = new WebAssembly.Instance(wasmModule, env.imports); |
| } |
| env.start(wasmInstance); |
| this.env = env; |
| this.lib = new FFILibrary(wasmInstance, env.imports); |
| this.memory = this.lib.memory; |
| this.exports = this.lib.exports; |
| this.objFactory = new Map(); |
| this.ctx = new RuntimeContext((name) => { |
| const autoAttachToScope = false; |
| |
| return this.getGlobalFuncInternal(name, autoAttachToScope); |
| }); |
| this.registerEnvGlobalPackedFuncs(); |
| this.registerObjectFactoryFuncs(); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| benchmark(run, dev, number = 10, repeat = 1) { |
| return __awaiter(this, void 0, void 0, function* () { |
| |
| const perf = compact.getPerformance(); |
| const results = []; |
| |
| this.withNewScope(run); |
| yield dev.sync(); |
| for (let k = 0; k < repeat; ++k) { |
| const tstart = perf.now(); |
| for (let i = 0; i < number; ++i) { |
| this.withNewScope(run); |
| } |
| yield dev.sync(); |
| const tend = perf.now(); |
| results.push((tend - tstart) / number); |
| } |
| return results; |
| }); |
| } |
| dispose() { |
| |
| |
| this.ctx.dispose(); |
| this.lib.dispose(); |
| } |
| |
| |
| |
| runtimeStatsText() { |
| if (this.lib.webGPUContext !== undefined) { |
| return this.lib.webGPUContext.runtimeStatsText(); |
| } |
| else { |
| return ""; |
| } |
| } |
| |
| |
| |
| beginScope() { |
| this.ctx.beginScope(); |
| } |
| |
| |
| |
| |
| |
| |
| |
| endScope() { |
| this.ctx.endScope(); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| withNewScope(action) { |
| this.beginScope(); |
| const val = action(); |
| this.endScope(); |
| return val; |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| attachToCurrentScope(obj) { |
| return this.ctx.attachToCurrentScope(obj); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| moveToParentScope(obj) { |
| return this.ctx.moveToParentScope(obj); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| detachFromCurrentScope(obj) { |
| return this.ctx.detachFromCurrentScope(obj); |
| } |
| |
| |
| |
| |
| |
| systemLib() { |
| return this.ctx.getSysLib(); |
| } |
| |
| |
| |
| |
| listGlobalFuncNames() { |
| const stack = this.lib.getOrAllocCallStack(); |
| const outSizeOffset = stack.allocPtrArray(2); |
| const outSizePtr = stack.ptrFromOffset(outSizeOffset); |
| const outArrayPtr = stack.ptrFromOffset(outSizeOffset + this.lib.sizeofPtr()); |
| this.lib.checkCall(this.exports.TVMFuncListGlobalNames(outSizePtr, outArrayPtr)); |
| const size = this.memory.loadI32(outSizePtr); |
| const array = this.memory.loadPointer(outArrayPtr); |
| const names = []; |
| for (let i = 0; i < size; ++i) { |
| names.push(this.memory.loadCString(this.memory.loadPointer(array + this.lib.sizeofPtr() * i))); |
| } |
| this.lib.recycleCallStack(stack); |
| return names; |
| } |
| |
| |
| |
| |
| |
| |
| registerFunc(name, func, override = false) { |
| this.withNewScope(() => { |
| const autoAttachToScope = true; |
| |
| const packedFunc = this.toPackedFuncInternal(func, autoAttachToScope); |
| const ioverride = override ? 1 : 0; |
| const stack = this.lib.getOrAllocCallStack(); |
| const nameOffset = stack.allocRawBytes(name.length + 1); |
| stack.storeRawBytes(nameOffset, support.StringToUint8Array(name)); |
| stack.commitToWasmMemory(); |
| this.lib.checkCall(this.lib.exports.TVMFuncRegisterGlobal(stack.ptrFromOffset(nameOffset), packedFunc._tvmPackedCell.getHandle(), ioverride)); |
| this.lib.recycleCallStack(stack); |
| }); |
| } |
| |
| |
| |
| |
| |
| |
| getGlobalFunc(name) { |
| return this.getGlobalFuncInternal(name, true); |
| } |
| getGlobalFuncInternal(name, autoAttachToScope = true) { |
| const stack = this.lib.getOrAllocCallStack(); |
| const nameOffset = stack.allocRawBytes(name.length + 1); |
| stack.storeRawBytes(nameOffset, support.StringToUint8Array(name)); |
| const outOffset = stack.allocPtrArray(1); |
| const outPtr = stack.ptrFromOffset(outOffset); |
| stack.commitToWasmMemory(outOffset); |
| this.lib.checkCall(this.exports.TVMFuncGetGlobal(stack.ptrFromOffset(nameOffset), outPtr)); |
| const handle = this.memory.loadPointer(outPtr); |
| this.lib.recycleCallStack(stack); |
| if (handle == 0) { |
| throw Error("Cannot find global function " + name); |
| } |
| const ret = this.makePackedFunc(handle); |
| if (autoAttachToScope) |
| this.ctx.attachToCurrentScope(ret); |
| return ret; |
| } |
| |
| |
| |
| |
| |
| |
| isPackedFunc(func) { |
| |
| return typeof func == "function" && func.hasOwnProperty("_tvmPackedCell"); |
| } |
| |
| |
| |
| |
| |
| |
| toPackedFunc(func) { |
| return this.toPackedFuncInternal(func, true); |
| } |
| toPackedFuncInternal(func, autoAttachToScope) { |
| if (this.isPackedFunc(func)) |
| return func; |
| const ret = this.createPackedFuncFromCFunc(this.wrapJSFuncAsPackedCFunc(func)); |
| if (autoAttachToScope) |
| return this.ctx.attachToCurrentScope(ret); |
| return ret; |
| } |
| |
| |
| |
| |
| |
| |
| createVirtualMachine(dev) { |
| const mod = this.ctx.detachFromCurrentScope(this.systemLib().getFunction("vm_load_executable")()); |
| return this.ctx.attachToCurrentScope(new VirtualMachine(mod, dev)); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| registerInitProgressCallback(cb) { |
| this.initProgressCallback.push(cb); |
| } |
| |
| |
| |
| |
| |
| |
| |
| getParamsFromCache(prefix, numParams) { |
| return this.ctx.paramModuleFromCache(prefix, new Scalar(numParams, "int32")).getFunction("get_params")(); |
| } |
| |
| |
| |
| |
| |
| ndarrayCacheGet(name) { |
| return this.ctx.arrayCacheGet(name); |
| } |
| |
| |
| |
| |
| |
| ndarrayCacheRemove(name) { |
| return this.ctx.arrayCacheRemove(name); |
| } |
| |
| |
| |
| |
| |
| ndarrayCacheUpdate(name, arr, override = false) { |
| this.ctx.arrayCacheUpdate(name, arr, this.scalar(override ? 1 : 0, "int32")); |
| } |
| |
| |
| |
| |
| |
| ndarrayCacheClear() { |
| this.ctx.arrayCacheClear(); |
| } |
| |
| |
| |
| |
| |
| |
| |
| fetchNDArrayCache(ndarrayCacheUrl, device) { |
| return __awaiter(this, void 0, void 0, function* () { |
| const cacheExists = yield caches.has("tvmjs"); |
| const jsonUrl = cacheExists |
| ? "/lib/WebLLM/vicuna-7b/ndarray-cache.json" |
| : new URL("ndarray-cache.json", ndarrayCacheUrl).href; |
| var list; |
| try { |
| list = yield (yield fetch(jsonUrl)).json(); |
| } |
| catch (err) { |
| this.env.logger("Cannot fetch " + jsonUrl); |
| } |
| yield this.fetchNDArrayCacheInternal(ndarrayCacheUrl, list["records"], device); |
| this.cacheMetadata = Object.assign(Object.assign({}, this.cacheMetadata), list["metadata"]); |
| }); |
| } |
| |
| |
| |
| |
| |
| |
| |
| fetchNDArrayCacheInternal(ndarrayCacheUrl, list, device) { |
| return __awaiter(this, void 0, void 0, function* () { |
| const perf = compact.getPerformance(); |
| let tstart = perf.now(); |
| let totalBytes = 0; |
| for (let i = 0; i < list.length; ++i) { |
| totalBytes += list[i].nbytes; |
| } |
| let fetchedBytes = 0; |
| let timeElapsed = 0; |
| const reportCallback = (iter) => { |
| |
| for (let j = 0; j < this.initProgressCallback.length; ++j) { |
| let text = "[" + iter + "/" + list.length + "]: "; |
| text += Math.ceil(fetchedBytes / (1024 * 1024)).toString() + " MB, "; |
| text += Math.floor(fetchedBytes * 100 / totalBytes).toString() + "% complete, "; |
| text += timeElapsed + " secs"; |
| this.initProgressCallback[j]({ |
| progress: fetchedBytes / totalBytes, |
| timeElapsed: timeElapsed, |
| text: text |
| }); |
| } |
| }; |
| for (let j = 0; j < this.initProgressCallback.length; ++j) { |
| this.initProgressCallback[j]({ |
| progress: fetchedBytes / totalBytes, |
| timeElapsed: 0, |
| text: "Start to fetch params", |
| }); |
| } |
| const cache = yield caches.open("tvmjs"); |
| for (let i = 0; i < list.length; ++i) { |
| reportCallback(i); |
| fetchedBytes += list[i].nbytes; |
| const dataUrl = new URL(list[i].dataPath, ndarrayCacheUrl).href; |
| const request = new Request(dataUrl); |
| let buffer; |
| try { |
| |
| let result = yield cache.match(request); |
| if (result === undefined) { |
| yield cache.add(request); |
| result = yield cache.match(request); |
| } |
| if (result == undefined) { |
| this.env.logger("Error: Cannot cache " + dataUrl + ", reloading will be slow"); |
| result = yield fetch(request); |
| } |
| buffer = yield result.arrayBuffer(); |
| } |
| catch (err) { |
| this.env.logger("Error: Cannot fetch " + dataUrl + " err= " + err); |
| throw err; |
| } |
| const shardRecords = list[i].records; |
| for (let j = 0; j < shardRecords.length; ++j) { |
| const rec = shardRecords[j]; |
| const cpu_arr = this.withNewScope(() => { |
| return this.detachFromCurrentScope(this.empty(rec.shape, rec.dtype, this.cpu())); |
| }); |
| const recSource = buffer.slice(rec.byteOffset, rec.byteOffset + rec.nbytes); |
| |
| this.ctx.arrayDecodeStorage(cpu_arr, new Uint8Array(recSource), rec.format); |
| |
| if (device.deviceType == DeviceStrToEnum.cpu) { |
| this.ndarrayCacheUpdate(rec.name, cpu_arr, false); |
| cpu_arr.dispose(); |
| } |
| else { |
| |
| const gpu_arr = this.withNewScope(() => { |
| return this.detachFromCurrentScope(this.empty(rec.shape, rec.dtype, device)); |
| }); |
| gpu_arr.copyFrom(cpu_arr); |
| yield device.sync(); |
| this.ndarrayCacheUpdate(rec.name, gpu_arr, false); |
| cpu_arr.dispose(); |
| gpu_arr.dispose(); |
| } |
| } |
| timeElapsed = Math.ceil((perf.now() - tstart) / 1000); |
| } |
| reportCallback(list.length); |
| }); |
| } |
| |
| |
| |
| |
| |
| |
| toDLDataType(dtype) { |
| if (dtype instanceof DLDataType) |
| return dtype; |
| if (typeof dtype == "string") { |
| let pattern = dtype; |
| let code, bits = 32, lanes = 1; |
| if (pattern.substring(0, 5) == "float") { |
| pattern = pattern.substring(5, pattern.length); |
| code = 2 ; |
| } |
| else if (pattern.substring(0, 3) == "int") { |
| pattern = pattern.substring(3, pattern.length); |
| code = 0 ; |
| } |
| else if (pattern.substring(0, 4) == "uint") { |
| pattern = pattern.substring(4, pattern.length); |
| code = 1 ; |
| } |
| else if (pattern.substring(0, 6) == "handle") { |
| pattern = pattern.substring(5, pattern.length); |
| code = 3 ; |
| bits = 64; |
| } |
| else { |
| throw new Error("Unknown dtype " + dtype); |
| } |
| const arr = pattern.split("x"); |
| if (arr.length >= 1) { |
| const parsed = parseInt(arr[0]); |
| if (parsed + "" == arr[0]) { |
| bits = parsed; |
| } |
| } |
| if (arr.length >= 2) { |
| lanes = parseInt(arr[1]); |
| } |
| return new DLDataType(code, bits, lanes); |
| } |
| else { |
| throw new Error("Unknown dtype " + dtype); |
| } |
| } |
| |
| |
| |
| |
| |
| |
| scalar(value, dtype) { |
| return new Scalar(value, dtype); |
| } |
| |
| |
| |
| |
| |
| |
| device(deviceType, deviceId = 0) { |
| return new DLDevice(deviceType, deviceId, this.lib); |
| } |
| |
| |
| |
| |
| cpu(deviceId = 0) { |
| return this.device("cpu", deviceId); |
| } |
| |
| |
| |
| |
| webgpu(deviceId = 0) { |
| return this.device("webgpu", deviceId); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| empty(shape, dtype = "float32", dev = this.device("cpu", 0)) { |
| dtype = this.toDLDataType(dtype); |
| shape = typeof shape == "number" ? [shape] : shape; |
| const stack = this.lib.getOrAllocCallStack(); |
| const shapeOffset = stack.allocRawBytes(shape.length * 8 ); |
| for (let i = 0; i < shape.length; ++i) { |
| stack.storeI64(shapeOffset + i * 8 , shape[i]); |
| } |
| const outOffset = stack.allocPtrArray(1); |
| const outPtr = stack.ptrFromOffset(outOffset); |
| stack.commitToWasmMemory(outOffset); |
| this.lib.checkCall(this.exports.TVMArrayAlloc(stack.ptrFromOffset(shapeOffset), shape.length, dtype.code, dtype.bits, dtype.lanes, dev.deviceType, dev.deviceId, outPtr)); |
| const ret = this.ctx.attachToCurrentScope(new NDArray(this.memory.loadPointer(outPtr), false, this.lib, this.ctx)); |
| this.lib.recycleCallStack(stack); |
| return ret; |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| uniform(shape, low, high, dev) { |
| const ret = this.empty(shape, "float32", dev); |
| const size = shape.reduce((a, b) => { |
| return a * b; |
| }, 1); |
| const scale = high - low; |
| const input = new Float32Array(size); |
| for (let i = 0; i < input.length; ++i) { |
| input[i] = low + Math.random() * scale; |
| } |
| return ret.copyFrom(input); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| sampleTopPFromLogits(logits, temperature, top_p) { |
| return this.ctx.sampleTopPFromLogits(logits, temperature, top_p, Math.random()); |
| } |
| |
| |
| |
| |
| bindCanvas(canvas) { |
| var _a; |
| (_a = this.lib.webGPUContext) === null || _a === void 0 ? void 0 : _a.bindCanvas(canvas); |
| } |
| |
| |
| |
| |
| |
| showImage(dataRGBA) { |
| var _a; |
| if (dataRGBA.shape.length != 2) { |
| throw Error("Require a height x width uint32 NDArray in RGBA" + |
| "get shape=" + dataRGBA.shape.toString() + " instead."); |
| } |
| if (dataRGBA.device.deviceType != DeviceStrToEnum.webgpu) { |
| throw new Error("Can only run showImage on WebGPU array, " + |
| "get " + DeviceEnumToStr[dataRGBA.device.deviceType] + " instead."); |
| } |
| if (dataRGBA.dtype != "uint32") { |
| throw Error("Require a height x width uint32 NDArray in RGBA, " + |
| "get " + dataRGBA.dtype + " instead."); |
| } |
| (_a = this.lib.webGPUContext) === null || _a === void 0 ? void 0 : _a.drawImageFromBuffer(dataRGBA.getDataPtr(), dataRGBA.shape[0], dataRGBA.shape[1]); |
| } |
| |
| |
| |
| clearCanvas() { |
| var _a; |
| (_a = this.lib.webGPUContext) === null || _a === void 0 ? void 0 : _a.clearCanvas(); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| makeTVMArray(inputs) { |
| return this.ctx.arrayMake(...inputs); |
| } |
| |
| |
| |
| |
| |
| makeShapeTuple(shape) { |
| const shapeArray = shape.map((value) => new Scalar(value, "int")); |
| return this.ctx.makeShapeTuple(...shapeArray); |
| } |
| |
| |
| |
| |
| |
| typeKey2Index(typeKey) { |
| const stack = this.lib.getOrAllocCallStack(); |
| const typeKeyOffset = stack.allocRawBytes(typeKey.length + 1); |
| stack.storeRawBytes(typeKeyOffset, support.StringToUint8Array(typeKey)); |
| const outOffset = stack.allocPtrArray(1); |
| const outPtr = stack.ptrFromOffset(outOffset); |
| stack.commitToWasmMemory(outOffset); |
| this.lib.checkCall(this.lib.exports.TVMObjectTypeKey2Index(stack.ptrFromOffset(typeKeyOffset), outPtr)); |
| const typeIndex = this.memory.loadU32(outPtr); |
| this.lib.recycleCallStack(stack); |
| return typeIndex; |
| } |
| |
| |
| |
| |
| |
| |
| registerObjectConstructor(typeKey, func, override = false) { |
| const typeIndex = this.typeKey2Index(typeKey); |
| if (this.objFactory.has(typeIndex)) { |
| if (!override) { |
| throw new Error("Type " + typeKey + " already registered"); |
| } |
| } |
| this.objFactory.set(typeIndex, func); |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| registerAsyncServerFunc(name, func, override = false) { |
| const asyncVariant = (...args) => { |
| const fargs = args.slice(0, args.length - 1); |
| |
| const callback = this.detachFromCurrentScope(args[args.length - 1]); |
| const promise = func(...fargs); |
| promise.then((rv) => { |
| callback(this.scalar(4 , "int32"), rv); |
| callback.dispose(); |
| }); |
| }; |
| this.registerFunc("__async." + name, asyncVariant, override); |
| } |
| |
| |
| |
| |
| asyncLoadWebGPUPiplines(mod) { |
| return __awaiter(this, void 0, void 0, function* () { |
| if (this.lib.webGPUContext == undefined) |
| throw Error("WebGPU not initialied"); |
| const webgpuContext = this.lib.webGPUContext; |
| this.beginScope(); |
| const fmap_str = mod.getFunction("webgpu.get_fmap", true)(); |
| let fmap = JSON.parse(fmap_str); |
| fmap.length; |
| const fGetShader = this.detachFromCurrentScope(mod.getFunction("webgpu.get_shader")); |
| const fUpdatePrebuild = this.detachFromCurrentScope(mod.getFunction("webgpu.update_prebuild")); |
| this.endScope(); |
| const perf = compact.getPerformance(); |
| const tstart = perf.now(); |
| let tlastReport = tstart; |
| let finishCounter = 0; |
| const fmapEntries = Object.entries(fmap); |
| let allEvents = Promise.resolve(); |
| for (const [key, finfo] of fmapEntries) { |
| const code = fGetShader(key); |
| support.assert(key == finfo.name); |
| const event = webgpuContext.createShaderAsync(finfo, code).then((func) => { |
| this.beginScope(); |
| fUpdatePrebuild(key, func); |
| this.endScope(); |
| }).then(() => { |
| finishCounter += 1; |
| const tend = perf.now(); |
| |
| if ((tend - tlastReport) < 1000 && finishCounter != fmapEntries.length) { |
| return; |
| } |
| tlastReport = tend; |
| const timeElapsed = Math.ceil((perf.now() - tstart) / 1000); |
| |
| for (let j = 0; j < this.initProgressCallback.length; ++j) { |
| const progress = finishCounter / fmapEntries.length; |
| let text = "Loading GPU shader modules[" + finishCounter + "/" + fmapEntries.length + "]: "; |
| text += Math.floor(progress * 100).toString() + "% completed, "; |
| text += timeElapsed + " secs elapsed."; |
| this.initProgressCallback[j]({ |
| progress: progress, |
| timeElapsed: timeElapsed, |
| text: text |
| }); |
| } |
| }); |
| allEvents = Promise.all([allEvents, event]).then(() => { }); |
| } |
| yield allEvents; |
| support.assert(finishCounter == fmapEntries.length); |
| }); |
| } |
| |
| |
| |
| |
| initWebGPU(device) { |
| const webGPUContext = new webgpu.WebGPUContext(this.memory, device); |
| this.registerFunc("wasm.WebGPUDeviceAPI", (name) => { |
| return webGPUContext.getDeviceAPI(name); |
| }); |
| this.registerFunc("wasm.WebGPUCreateShader", (info, code) => { |
| const finfo = JSON.parse(info); |
| return webGPUContext.createShader(finfo, code); |
| }); |
| this.registerAsyncServerFunc("wasm.WebGPUWaitForTasks", () => __awaiter(this, void 0, void 0, function* () { |
| yield webGPUContext.sync(); |
| })); |
| this.lib.webGPUContext = webGPUContext; |
| } |
| |
| registerObjectFactoryFuncs() { |
| this.registerObjectConstructor("Array", (handle, lib, ctx) => { |
| return new TVMArray(handle, lib, ctx); |
| }); |
| } |
| |
| registerEnvGlobalPackedFuncs() { |
| |
| const perf = compact.getPerformance(); |
| |
| const timeExecution = (finvoke, dev, nstep, repeat, minRepeatMs, limitZeroTimeIterations, cooldownIntervalMs, repeatsToCooldown) => __awaiter(this, void 0, void 0, function* () { |
| |
| |
| |
| this.ctx.detachFromCurrentScope(finvoke); |
| finvoke(this.scalar(1, "int32")); |
| yield dev.sync(); |
| const result = []; |
| let setupNumber = nstep; |
| for (let i = 0; i < repeat; ++i) { |
| let durationMs = 0.0; |
| let absoluteZeroTimes = 0; |
| do { |
| if (durationMs > 0.0) { |
| let golden_ratio = 1.618; |
| setupNumber = Math.floor(Math.max(minRepeatMs / (durationMs / setupNumber) + 1, setupNumber * golden_ratio)); |
| } |
| const tstart = perf.now(); |
| finvoke(this.scalar(setupNumber, "int32")); |
| yield dev.sync(); |
| const tend = perf.now(); |
| durationMs = tend - tstart; |
| if (durationMs == 0) { |
| absoluteZeroTimes++; |
| } |
| } while (durationMs < minRepeatMs && absoluteZeroTimes < limitZeroTimeIterations); |
| const speed = durationMs / setupNumber / 1000; |
| result.push(speed); |
| if (cooldownIntervalMs > 0.0 && (i % repeatsToCooldown) == 0) { |
| yield new Promise(r => setTimeout(r, cooldownIntervalMs)); |
| } |
| } |
| const ret = new Float64Array(result.length); |
| ret.set(result); |
| |
| finvoke.dispose(); |
| return new Uint8Array(ret.buffer); |
| }); |
| const addOne = (x) => __awaiter(this, void 0, void 0, function* () { |
| yield new Promise(resolve => setTimeout(resolve, 100)); |
| return x + 1; |
| }); |
| this.registerAsyncServerFunc("wasm.TimeExecution", timeExecution); |
| this.registerAsyncServerFunc("testing.asyncAddOne", addOne); |
| } |
| createPackedFuncFromCFunc(func) { |
| let findex = this.env.packedCFuncTable.length; |
| if (this.env.packedCFuncTableFreeId.length != 0) { |
| findex = this.env.packedCFuncTableFreeId.pop(); |
| } |
| else { |
| this.env.packedCFuncTable.push(undefined); |
| } |
| this.env.packedCFuncTable[findex] = func; |
| const stack = this.lib.getOrAllocCallStack(); |
| const outOffset = stack.allocPtrArray(1); |
| const outPtr = stack.ptrFromOffset(outOffset); |
| this.lib.checkCall(this.exports |
| .TVMWasmFuncCreateFromCFunc(findex, outPtr)); |
| const ret = this.makePackedFunc(this.memory.loadPointer(outPtr)); |
| this.lib.recycleCallStack(stack); |
| return ret; |
| } |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| setPackedArguments(stack, args, argsValue, argsCode) { |
| for (let i = 0; i < args.length; ++i) { |
| let val = args[i]; |
| const tp = typeof val; |
| const valueOffset = argsValue + i * 8 ; |
| const codeOffset = argsCode + i * 4 ; |
| if (val instanceof NDArray) { |
| if (!val.isView) { |
| stack.storePtr(valueOffset, val.getHandle()); |
| stack.storeI32(codeOffset, 13 ); |
| } |
| else { |
| stack.storePtr(valueOffset, val.getHandle()); |
| stack.storeI32(codeOffset, 7 ); |
| } |
| } |
| else if (val instanceof Scalar) { |
| if (val.dtype.startsWith("int") || val.dtype.startsWith("uint")) { |
| stack.storeI64(valueOffset, val.value); |
| stack.storeI32(codeOffset, 0 ); |
| } |
| else if (val.dtype.startsWith("float")) { |
| stack.storeF64(valueOffset, val.value); |
| stack.storeI32(codeOffset, 2 ); |
| } |
| else { |
| support.assert(val.dtype == "handle", "Expect handle"); |
| stack.storePtr(valueOffset, val.value); |
| stack.storeI32(codeOffset, 3 ); |
| } |
| } |
| else if (val instanceof DLDevice) { |
| stack.storeI32(valueOffset, val.deviceType); |
| stack.storeI32(valueOffset + 4 , val.deviceType); |
| stack.storeI32(codeOffset, 6 ); |
| } |
| else if (tp == "number") { |
| stack.storeF64(valueOffset, val); |
| stack.storeI32(codeOffset, 2 ); |
| |
| } |
| else if (tp == "function" && val.hasOwnProperty("_tvmPackedCell")) { |
| stack.storePtr(valueOffset, val._tvmPackedCell.getHandle()); |
| stack.storeI32(codeOffset, 10 ); |
| } |
| else if (val === null || val == undefined) { |
| stack.storePtr(valueOffset, 0); |
| stack.storeI32(codeOffset, 4 ); |
| } |
| else if (tp == "string") { |
| stack.allocThenSetArgString(valueOffset, val); |
| stack.storeI32(codeOffset, 11 ); |
| } |
| else if (val instanceof Uint8Array) { |
| stack.allocThenSetArgBytes(valueOffset, val); |
| stack.storeI32(codeOffset, 12 ); |
| } |
| else if (val instanceof Function) { |
| val = this.toPackedFuncInternal(val, false); |
| stack.tempArgs.push(val); |
| stack.storePtr(valueOffset, val._tvmPackedCell.getHandle()); |
| stack.storeI32(codeOffset, 10 ); |
| } |
| else if (val instanceof Module) { |
| stack.storePtr(valueOffset, val.getHandle()); |
| stack.storeI32(codeOffset, 9 ); |
| } |
| else if (val instanceof TVMObject) { |
| stack.storePtr(valueOffset, val.getHandle()); |
| stack.storeI32(codeOffset, 8 ); |
| } |
| else { |
| throw new Error("Unsupported argument type " + tp); |
| } |
| } |
| } |
| wrapJSFuncAsPackedCFunc(func) { |
| const lib = this.lib; |
| return (argValues, argCodes, nargs, ret, |
| |
| _handle) => { |
| const jsArgs = []; |
| |
| this.ctx.beginScope(); |
| for (let i = 0; i < nargs; ++i) { |
| const valuePtr = argValues + i * 8 ; |
| const codePtr = argCodes + i * 4 ; |
| let tcode = lib.memory.loadI32(codePtr); |
| if (tcode == 8 || |
| tcode == 14 || |
| tcode == 10 || |
| tcode == 13 || |
| tcode == 9 ) { |
| lib.checkCall(lib.exports.TVMCbArgToReturn(valuePtr, codePtr)); |
| } |
| tcode = lib.memory.loadI32(codePtr); |
| jsArgs.push(this.retValueToJS(valuePtr, tcode, true)); |
| } |
| const rv = func(...jsArgs); |
| |
| this.ctx.endScope(); |
| if (rv !== undefined && rv !== null) { |
| const stack = lib.getOrAllocCallStack(); |
| const valueOffset = stack.allocRawBytes(8 ); |
| const codeOffset = stack.allocRawBytes(4 ); |
| this.setPackedArguments(stack, [rv], valueOffset, codeOffset); |
| const valuePtr = stack.ptrFromOffset(valueOffset); |
| const codePtr = stack.ptrFromOffset(codeOffset); |
| stack.commitToWasmMemory(); |
| lib.checkCall(lib.exports.TVMCFuncSetReturn(ret, valuePtr, codePtr, 1)); |
| lib.recycleCallStack(stack); |
| } |
| return 0; |
| }; |
| } |
| makePackedFunc(handle) { |
| const cell = new PackedFuncCell(handle, this.lib); |
| const packedFunc = (...args) => { |
| const stack = this.lib.getOrAllocCallStack(); |
| const valueOffset = stack.allocRawBytes(8 * args.length); |
| const tcodeOffset = stack.allocRawBytes(4 * args.length); |
| this.setPackedArguments(stack, args, valueOffset, tcodeOffset); |
| const rvalueOffset = stack.allocRawBytes(8 ); |
| const rcodeOffset = stack.allocRawBytes(4 ); |
| const rvaluePtr = stack.ptrFromOffset(rvalueOffset); |
| const rcodePtr = stack.ptrFromOffset(rcodeOffset); |
| |
| stack.commitToWasmMemory(rvalueOffset); |
| this.lib.checkCall(this.exports.TVMFuncCall(cell.getHandle(), stack.ptrFromOffset(valueOffset), stack.ptrFromOffset(tcodeOffset), args.length, rvaluePtr, rcodePtr)); |
| const ret = this.retValueToJS(rvaluePtr, this.memory.loadI32(rcodePtr), false); |
| this.lib.recycleCallStack(stack); |
| return ret; |
| }; |
| |
| |
| const ret = packedFunc; |
| ret.dispose = () => { |
| cell.dispose(); |
| }; |
| ret._tvmPackedCell = cell; |
| return ret; |
| } |
| |
| |
| |
| |
| |
| |
| |
| retValueToJS(rvaluePtr, tcode, callbackArg) { |
| switch (tcode) { |
| case 0 : |
| case 1 : |
| return this.memory.loadI64(rvaluePtr); |
| case 2 : |
| return this.memory.loadF64(rvaluePtr); |
| case 3 : { |
| return this.memory.loadPointer(rvaluePtr); |
| } |
| case 13 : { |
| return this.ctx.attachToCurrentScope(new NDArray(this.memory.loadPointer(rvaluePtr), false, this.lib, this.ctx)); |
| } |
| case 7 : { |
| support.assert(callbackArg); |
| |
| return new NDArray(this.memory.loadPointer(rvaluePtr), true, this.lib, this.ctx); |
| } |
| case 10 : { |
| return this.ctx.attachToCurrentScope(this.makePackedFunc(this.memory.loadPointer(rvaluePtr))); |
| } |
| case 9 : { |
| return this.ctx.attachToCurrentScope(new Module(this.memory.loadPointer(rvaluePtr), this.lib, (ptr) => { |
| return this.ctx.attachToCurrentScope(this.makePackedFunc(ptr)); |
| })); |
| } |
| case 8 : { |
| const obj = new TVMObject(this.memory.loadPointer(rvaluePtr), this.lib, this.ctx); |
| const func = this.objFactory.get(obj.typeIndex()); |
| if (func != undefined) { |
| return this.ctx.attachToCurrentScope(func(obj.getHandle(), this.lib, this.ctx)); |
| } |
| else { |
| return this.ctx.attachToCurrentScope(obj); |
| } |
| } |
| case 4 : return undefined; |
| case 6 : { |
| const deviceType = this.memory.loadI32(rvaluePtr); |
| const deviceId = this.memory.loadI32(rvaluePtr + 4 ); |
| return this.device(deviceType, deviceId); |
| } |
| case 11 : { |
| const ret = this.memory.loadCString(this.memory.loadPointer(rvaluePtr)); |
| return ret; |
| } |
| case 12 : { |
| return this.memory.loadTVMBytes(this.memory.loadPointer(rvaluePtr)); |
| } |
| default: |
| throw new Error("Unsupported return type code=" + tcode); |
| } |
| } |
| } |
| exports.Instance = Instance; |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| function instantiate(bufferSource, importObject = {}, logger = console.log) { |
| const env = new environment.Environment(importObject, logger); |
| return WebAssembly.instantiate(bufferSource, env.imports).then((result) => { |
| return new Instance(result.module, {}, result.instance, env); |
| }); |
| } |
| exports.instantiate = instantiate; |
|
|
| }); |
|
|
| unwrapExports(runtime); |
| runtime.instantiate; |
| runtime.Instance; |
| runtime.VirtualMachine; |
| runtime.TVMArray; |
| runtime.TVMObject; |
| runtime.Module; |
| runtime.NDArray; |
| runtime.DLDataType; |
| runtime.DLDevice; |
| runtime.Scalar; |
|
|
| var rpc_server = createCommonjsModule(function (module, exports) { |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| var __awaiter = (commonjsGlobal && commonjsGlobal.__awaiter) || function (thisArg, _arguments, P, generator) { |
| function adopt(value) { return value instanceof P ? value : new P(function (resolve) { resolve(value); }); } |
| return new (P || (P = Promise))(function (resolve, reject) { |
| function fulfilled(value) { try { step(generator.next(value)); } catch (e) { reject(e); } } |
| function rejected(value) { try { step(generator["throw"](value)); } catch (e) { reject(e); } } |
| function step(result) { result.done ? resolve(result.value) : adopt(result.value).then(fulfilled, rejected); } |
| step((generator = generator.apply(thisArg, _arguments || [])).next()); |
| }); |
| }; |
| Object.defineProperty(exports, "__esModule", { value: true }); |
| exports.RPCServer = void 0; |
|
|
|
|
|
|
|
|
| var RPCServerState; |
| (function (RPCServerState) { |
| RPCServerState[RPCServerState["InitHeader"] = 0] = "InitHeader"; |
| RPCServerState[RPCServerState["InitHeaderKey"] = 1] = "InitHeaderKey"; |
| RPCServerState[RPCServerState["InitServer"] = 2] = "InitServer"; |
| RPCServerState[RPCServerState["WaitForCallback"] = 3] = "WaitForCallback"; |
| RPCServerState[RPCServerState["ReceivePacketHeader"] = 4] = "ReceivePacketHeader"; |
| RPCServerState[RPCServerState["ReceivePacketBody"] = 5] = "ReceivePacketBody"; |
| })(RPCServerState || (RPCServerState = {})); |
| |
| const RPC_MAGIC = 0xff271; |
| |
| |
| |
| class ByteStreamReader { |
| constructor(bytes) { |
| this.offset = 0; |
| this.bytes = bytes; |
| } |
| readU32() { |
| const i = this.offset; |
| const b = this.bytes; |
| const val = b[i] | (b[i + 1] << 8) | (b[i + 2] << 16) | (b[i + 3] << 24); |
| this.offset += 4; |
| return val; |
| } |
| readU64() { |
| const val = this.readU32(); |
| this.offset += 4; |
| return val; |
| } |
| readByteArray() { |
| const len = this.readU64(); |
| support.assert(this.offset + len <= this.bytes.byteLength); |
| const ret = new Uint8Array(len); |
| ret.set(this.bytes.slice(this.offset, this.offset + len)); |
| this.offset += len; |
| return ret; |
| } |
| } |
| |
| |
| |
| class RPCServer { |
| constructor(url, key, getImports, logger = console.log, ndarrayCacheUrl = "", ndarrayCacheDevice = "cpu", initProgressCallback = undefined, asyncOnServerLoad = undefined) { |
| this.state = RPCServerState.InitHeader; |
| this.pendingSend = Promise.resolve(); |
| this.inst = undefined; |
| this.globalObjects = []; |
| this.currPacketLength = 0; |
| this.remoteKeyLength = 0; |
| this.pendingBytes = 0; |
| this.buffredBytes = 0; |
| this.messageQueue = []; |
| this.url = url; |
| this.key = key; |
| this.name = "WebSocketRPCServer[" + this.key + "]: "; |
| this.getImports = getImports; |
| this.logger = logger; |
| this.ndarrayCacheUrl = ndarrayCacheUrl; |
| this.ndarrayCacheDevice = ndarrayCacheDevice; |
| this.initProgressCallback = initProgressCallback; |
| this.asyncOnServerLoad = asyncOnServerLoad; |
| this.checkLittleEndian(); |
| this.socket = compact.createWebSocket(url); |
| this.socket.binaryType = "arraybuffer"; |
| this.socket.addEventListener("open", (event) => { |
| return this.onOpen(event); |
| }); |
| this.socket.addEventListener("message", (event) => { |
| return this.onMessage(event); |
| }); |
| this.socket.addEventListener("close", (event) => { |
| return this.onClose(event); |
| }); |
| } |
| |
| onClose(_event) { |
| if (this.inst !== undefined) { |
| this.globalObjects.forEach(obj => { |
| obj.dispose(); |
| }); |
| this.log(this.inst.runtimeStatsText()); |
| this.inst.dispose(); |
| } |
| if (this.state == RPCServerState.ReceivePacketHeader) { |
| this.log("Closing the server in clean state"); |
| this.log("Automatic reconnecting.."); |
| new RPCServer(this.url, this.key, this.getImports, this.logger, this.ndarrayCacheUrl, this.ndarrayCacheDevice, this.initProgressCallback, this.asyncOnServerLoad); |
| } |
| else { |
| this.log("Closing the server, final state=" + this.state); |
| } |
| } |
| |
| onOpen(_event) { |
| |
| let bkey = support.StringToUint8Array("server:" + this.key); |
| bkey = bkey.slice(0, bkey.length - 1); |
| const intbuf = new Int32Array(1); |
| intbuf[0] = RPC_MAGIC; |
| this.socket.send(intbuf); |
| intbuf[0] = bkey.length; |
| this.socket.send(intbuf); |
| this.socket.send(bkey); |
| this.log("connected..."); |
| |
| this.requestBytes(4 + 4 ); |
| this.state = RPCServerState.InitHeader; |
| } |
| |
| onMessage(event) { |
| const buffer = event.data; |
| this.buffredBytes += buffer.byteLength; |
| this.messageQueue.push(new Uint8Array(buffer)); |
| this.processEvents(); |
| } |
| |
| processEvents() { |
| while (this.buffredBytes >= this.pendingBytes && this.pendingBytes != 0) { |
| this.onDataReady(); |
| } |
| } |
| |
| onDataReady() { |
| switch (this.state) { |
| case RPCServerState.InitHeader: { |
| this.handleInitHeader(); |
| break; |
| } |
| case RPCServerState.InitHeaderKey: { |
| this.handleInitHeaderKey(); |
| break; |
| } |
| case RPCServerState.ReceivePacketHeader: { |
| this.currPacketHeader = this.readFromBuffer(8 ); |
| const reader = new ByteStreamReader(this.currPacketHeader); |
| this.currPacketLength = reader.readU64(); |
| support.assert(this.pendingBytes == 0); |
| this.requestBytes(this.currPacketLength); |
| this.state = RPCServerState.ReceivePacketBody; |
| break; |
| } |
| case RPCServerState.ReceivePacketBody: { |
| const body = this.readFromBuffer(this.currPacketLength); |
| support.assert(this.pendingBytes == 0); |
| support.assert(this.currPacketHeader !== undefined); |
| this.onPacketReady(this.currPacketHeader, body); |
| break; |
| } |
| case RPCServerState.WaitForCallback: { |
| support.assert(this.pendingBytes == 0); |
| break; |
| } |
| default: { |
| throw new Error("Cannot handle state " + this.state); |
| } |
| } |
| } |
| onPacketReady(header, body) { |
| if (this.inst === undefined) { |
| |
| const reader = new ByteStreamReader(body); |
| |
| reader.readU32(); |
| |
| support.Uint8ArrayToString(reader.readByteArray()); |
| const nargs = reader.readU32(); |
| const tcodes = []; |
| const args = []; |
| for (let i = 0; i < nargs; ++i) { |
| tcodes.push(reader.readU32()); |
| } |
| for (let i = 0; i < nargs; ++i) { |
| const tcode = tcodes[i]; |
| if (tcode == 11 ) { |
| const str = support.Uint8ArrayToString(reader.readByteArray()); |
| args.push(str); |
| } |
| else if (tcode == 12 ) { |
| args.push(reader.readByteArray()); |
| } |
| else { |
| throw new Error("cannot support type code " + tcode); |
| } |
| } |
| this.onInitServer(args, header, body); |
| } |
| else { |
| support.assert(this.serverRecvData !== undefined); |
| this.serverRecvData(header, body); |
| this.requestBytes(8 ); |
| this.state = RPCServerState.ReceivePacketHeader; |
| } |
| } |
| |
| onInitServer(args, header, body) { |
| |
| support.assert(args[0] == "rpc.WasmSession"); |
| support.assert(this.pendingBytes == 0); |
| const asyncInitServer = () => __awaiter(this, void 0, void 0, function* () { |
| support.assert(args[1] instanceof Uint8Array); |
| const inst = yield runtime.instantiate(args[1].buffer, this.getImports(), this.logger); |
| try { |
| const output = yield webgpu.detectGPUDevice(); |
| if (output !== undefined) { |
| const label = "WebGPU: " + output.adapterInfo.description; |
| this.log("Initialize GPU device: " + label); |
| inst.initWebGPU(output.device); |
| } |
| else { |
| this.log("Cannot find WebGPU device in the env"); |
| } |
| } |
| catch (err) { |
| this.log("Cannnot initialize WebGPU, " + err.toString()); |
| } |
| this.inst = inst; |
| |
| this.inst.beginScope(); |
| if (this.initProgressCallback !== undefined) { |
| this.inst.registerInitProgressCallback(this.initProgressCallback); |
| } |
| if (this.ndarrayCacheUrl.length != 0) { |
| if (this.ndarrayCacheDevice == "cpu") { |
| yield this.inst.fetchNDArrayCache(this.ndarrayCacheUrl, this.inst.cpu()); |
| } |
| else { |
| support.assert(this.ndarrayCacheDevice == "webgpu"); |
| yield this.inst.fetchNDArrayCache(this.ndarrayCacheUrl, this.inst.webgpu()); |
| } |
| } |
| support.assert(this.inst !== undefined); |
| if (this.asyncOnServerLoad !== undefined) { |
| yield this.asyncOnServerLoad(this.inst); |
| } |
| const fcreate = this.inst.getGlobalFunc("rpc.CreateEventDrivenServer"); |
| const messageHandler = fcreate((cbytes) => { |
| support.assert(this.inst !== undefined); |
| if (this.socket.readyState == 1) { |
| |
| |
| |
| const sendDataWithCongestionControl = () => __awaiter(this, void 0, void 0, function* () { |
| const packetSize = 4 << 10; |
| const maxBufferAmount = 4 * packetSize; |
| const waitTimeMs = 20; |
| for (let offset = 0; offset < cbytes.length; offset += packetSize) { |
| const end = Math.min(offset + packetSize, cbytes.length); |
| while (this.socket.bufferedAmount >= maxBufferAmount) { |
| yield new Promise((r) => setTimeout(r, waitTimeMs)); |
| } |
| this.socket.send(cbytes.slice(offset, end)); |
| } |
| }); |
| |
| this.pendingSend = this.pendingSend.then(sendDataWithCongestionControl); |
| |
| return this.inst.scalar(cbytes.length, "int32"); |
| } |
| else { |
| return this.inst.scalar(0, "int32"); |
| } |
| }, this.name, this.key); |
| |
| this.globalObjects.push(this.inst.detachFromCurrentScope(messageHandler)); |
| const writeFlag = this.inst.scalar(3, "int32"); |
| this.serverRecvData = (header, body) => { |
| if (messageHandler(header, writeFlag) == 0) { |
| this.socket.close(); |
| } |
| if (messageHandler(body, writeFlag) == 0) { |
| this.socket.close(); |
| } |
| }; |
| |
| |
| |
| |
| const flocal = this.inst.getGlobalFunc("wasm.LocalSession"); |
| const localSession = flocal(); |
| support.assert(localSession instanceof runtime.Module); |
| |
| this.inst.registerFunc("rpc.WasmSession", |
| |
| (_args) => { |
| return localSession; |
| }); |
| messageHandler(header, writeFlag); |
| messageHandler(body, writeFlag); |
| this.log("Finish initializing the Wasm Server.."); |
| this.requestBytes(8 ); |
| this.state = RPCServerState.ReceivePacketHeader; |
| |
| this.processEvents(); |
| |
| this.inst.endScope(); |
| }); |
| this.state = RPCServerState.WaitForCallback; |
| asyncInitServer(); |
| } |
| log(msg) { |
| this.logger(this.name + msg); |
| } |
| handleInitHeader() { |
| const reader = new ByteStreamReader(this.readFromBuffer(4 * 2)); |
| const magic = reader.readU32(); |
| if (magic == RPC_MAGIC + 1) { |
| throw new Error("key: " + this.key + " has already been used in proxy"); |
| } |
| else if (magic == RPC_MAGIC + 2) { |
| throw new Error("RPCProxy do not have matching client key " + this.key); |
| } |
| support.assert(magic == RPC_MAGIC, this.url + " is not an RPC Proxy"); |
| this.remoteKeyLength = reader.readU32(); |
| support.assert(this.pendingBytes == 0); |
| this.requestBytes(this.remoteKeyLength); |
| this.state = RPCServerState.InitHeaderKey; |
| } |
| handleInitHeaderKey() { |
| |
| support.Uint8ArrayToString(this.readFromBuffer(this.remoteKeyLength)); |
| support.assert(this.pendingBytes == 0); |
| this.requestBytes(8 ); |
| this.state = RPCServerState.ReceivePacketHeader; |
| } |
| checkLittleEndian() { |
| const a = new ArrayBuffer(4); |
| const b = new Uint8Array(a); |
| const c = new Uint32Array(a); |
| b[0] = 0x11; |
| b[1] = 0x22; |
| b[2] = 0x33; |
| b[3] = 0x44; |
| support.assert(c[0] === 0x44332211, "RPCServer little endian to work"); |
| } |
| requestBytes(nbytes) { |
| this.pendingBytes += nbytes; |
| } |
| readFromBuffer(nbytes) { |
| const ret = new Uint8Array(nbytes); |
| let ptr = 0; |
| while (ptr < nbytes) { |
| support.assert(this.messageQueue.length != 0); |
| const nleft = nbytes - ptr; |
| if (this.messageQueue[0].byteLength <= nleft) { |
| const buffer = this.messageQueue.shift(); |
| ret.set(buffer, ptr); |
| ptr += buffer.byteLength; |
| } |
| else { |
| const buffer = this.messageQueue[0]; |
| ret.set(buffer.slice(0, nleft), ptr); |
| this.messageQueue[0] = buffer.slice(nleft, buffer.byteLength); |
| ptr += nleft; |
| } |
| } |
| this.buffredBytes -= nbytes; |
| this.pendingBytes -= nbytes; |
| return ret; |
| } |
| } |
| exports.RPCServer = RPCServer; |
|
|
| }); |
|
|
| unwrapExports(rpc_server); |
| rpc_server.RPCServer; |
|
|
| var dist = createCommonjsModule(function (module, exports) { |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| |
| Object.defineProperty(exports, "__esModule", { value: true }); |
|
|
| Object.defineProperty(exports, "Scalar", { enumerable: true, get: function () { return runtime.Scalar; } }); |
| Object.defineProperty(exports, "DLDevice", { enumerable: true, get: function () { return runtime.DLDevice; } }); |
| Object.defineProperty(exports, "DLDataType", { enumerable: true, get: function () { return runtime.DLDataType; } }); |
| Object.defineProperty(exports, "Module", { enumerable: true, get: function () { return runtime.Module; } }); |
| Object.defineProperty(exports, "NDArray", { enumerable: true, get: function () { return runtime.NDArray; } }); |
| Object.defineProperty(exports, "TVMArray", { enumerable: true, get: function () { return runtime.TVMArray; } }); |
| Object.defineProperty(exports, "Instance", { enumerable: true, get: function () { return runtime.Instance; } }); |
| Object.defineProperty(exports, "instantiate", { enumerable: true, get: function () { return runtime.instantiate; } }); |
|
|
| Object.defineProperty(exports, "RPCServer", { enumerable: true, get: function () { return rpc_server.RPCServer; } }); |
|
|
| Object.defineProperty(exports, "wasmPath", { enumerable: true, get: function () { return support.wasmPath; } }); |
|
|
| Object.defineProperty(exports, "detectGPUDevice", { enumerable: true, get: function () { return webgpu.detectGPUDevice; } }); |
| var support_2 = support; |
| Object.defineProperty(exports, "assert", { enumerable: true, get: function () { return support_2.assert; } }); |
|
|
| }); |
|
|
| var index = unwrapExports(dist); |
| var dist_1 = dist.Scalar; |
| var dist_2 = dist.DLDevice; |
| var dist_3 = dist.DLDataType; |
| var dist_4 = dist.Module; |
| var dist_5 = dist.NDArray; |
| var dist_6 = dist.TVMArray; |
| var dist_7 = dist.Instance; |
| var dist_8 = dist.instantiate; |
| var dist_9 = dist.RPCServer; |
| var dist_10 = dist.wasmPath; |
| var dist_11 = dist.detectGPUDevice; |
| var dist_12 = dist.assert; |
|
|
| exports.DLDataType = dist_3; |
| exports.DLDevice = dist_2; |
| exports.Instance = dist_7; |
| exports.Module = dist_4; |
| exports.NDArray = dist_5; |
| exports.RPCServer = dist_9; |
| exports.Scalar = dist_1; |
| exports.TVMArray = dist_6; |
| exports.assert = dist_12; |
| exports["default"] = index; |
| exports.detectGPUDevice = dist_11; |
| exports.instantiate = dist_8; |
| exports.wasmPath = dist_10; |
|
|
| Object.defineProperty(exports, '__esModule', { value: true }); |
|
|
| })); |
|
|