mirror of
https://github.com/saymrwulf/onnxruntime.git
synced 2026-07-30 20:18:08 +00:00
Merge remote-tracking branch 'upstream/main' into snnn/update_pybind
This commit is contained in:
Changming Sun
commit
94d4827f69
62 changed files with 463 additions and 953 deletions
|
|
@ -1,30 +1,3 @@
|
|||
# CGManifest Files
|
||||
This directory contains CGManifest (cgmanifest.json) files.
|
||||
See [here](https://docs.opensource.microsoft.com/tools/cg/cgmanifest.html) for details.
|
||||
|
||||
## `cgmanifests/generated/cgmanifest.json`
|
||||
This file contains generated CGManifest entries.
|
||||
|
||||
It covers these dependencies:
|
||||
- git submodules
|
||||
- dependencies from the Dockerfile `tools/ci_build/github/linux/docker/Dockerfile.manylinux2014_cuda11`
|
||||
- the entries in [../cmake/deps.txt](../cmake/deps.txt)
|
||||
|
||||
If any of these dependencies change, this file should be updated.
|
||||
**When updating, please regenerate instead of editing manually.**
|
||||
|
||||
### How to Generate
|
||||
1. Change to the repository root directory.
|
||||
2. Ensure the git submodules are checked out and up to date. For example, with:
|
||||
```
|
||||
$ git submodule update --init --recursive
|
||||
```
|
||||
3. Run the generator script:
|
||||
```
|
||||
$ python cgmanifests/generate_cgmanifest.py --username <xxx> --token <your_access_token>
|
||||
```
|
||||
|
||||
Please supply your github username and access token to the script. If you don't have a token, you can generate one at https://github.com/settings/tokens. This is for authenticating with Github REST API so that you would not hit the rate limit.
|
||||
|
||||
## `cgmanifests/cgmanifest.json`
|
||||
This file contains non-generated CGManifest entries. Please edit directly as needed.
|
||||
See [here](https://docs.opensource.microsoft.com/tools/cg/cgmanifest.html) for details.
|
||||
|
|
@ -1,15 +1,6 @@
|
|||
{
|
||||
"$schema": "https://json.schemastore.org/component-detection-manifest.json",
|
||||
"Registrations": [
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "215105818dfde3174fe799600bb0f3cae233d0bf",
|
||||
"repositoryUrl": "https://github.com/abseil/abseil-cpp.git"
|
||||
}
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"Type": "maven",
|
||||
|
|
@ -131,16 +122,6 @@
|
|||
}
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "b31f58de6fa8bbda5353b3c77d9be4914399724d",
|
||||
"repositoryUrl": "https://github.com/pytorch/pytorch.git"
|
||||
},
|
||||
"comments": "pytorch 1.6 used by onnxruntime training image"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
|
|
@ -189,15 +170,6 @@
|
|||
}
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"git": {
|
||||
"commitHash": "02a2a458ac15912d7d87cc1171e811b0c5219ece",
|
||||
"repositoryUrl": "https://github.com/grpc/grpc"
|
||||
},
|
||||
"type": "git"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"git": {
|
||||
|
|
@ -324,116 +296,6 @@
|
|||
"comments": "For Nodejs binding"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "aead4d751c2101e23336aa73f2380df83e7a13f3",
|
||||
"repositoryUrl": "https://github.com/pypa/manylinux"
|
||||
},
|
||||
"comments": "For building our CI build docker image"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "c974557598645360fbabac71352b083117e3cc17",
|
||||
"repositoryUrl": "https://gitlab.kitware.com/cmake/cmake"
|
||||
},
|
||||
"comments": "CMake 3.24.3. For building our CI build docker image"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "1e5d33e9b9b8631b36f061103a30208b206fd03a",
|
||||
"repositoryUrl": "https://github.com/python/cpython"
|
||||
},
|
||||
"comments": "Python 3.9.1"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "6503f05dd59e26a9986bdea097b3da9b3546f45b",
|
||||
"repositoryUrl": "https://github.com/python/cpython"
|
||||
},
|
||||
"comments": "Python 3.8.7"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "13c94747c74437e594b7fc242ff7da668e81887c",
|
||||
"repositoryUrl": "https://github.com/python/cpython"
|
||||
},
|
||||
"comments": "Python 3.7.9"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "c0a9afe2ac1820409e6173bd1893ebee2cf50270",
|
||||
"repositoryUrl": "https://github.com/python/cpython"
|
||||
},
|
||||
"comments": "Python 3.6.12"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "426b022776672fdf3d71ddd98d89af341c88080f",
|
||||
"repositoryUrl": "https://github.com/python/cpython"
|
||||
},
|
||||
"comments": "Python 3.5.10"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "pip",
|
||||
"pip": {
|
||||
"Name": "transformers",
|
||||
"Version": "4.38.0"
|
||||
},
|
||||
"comments": "Installed in the training docker image"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "pip",
|
||||
"pip": {
|
||||
"Name": "msgpack",
|
||||
"Version": "1.0.0"
|
||||
},
|
||||
"comments": "Installed in the training docker image"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "pip",
|
||||
"pip": {
|
||||
"Name": "tensorboardX",
|
||||
"Version": "1.8"
|
||||
},
|
||||
"comments": "Installed in the training docker image"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "pip",
|
||||
"pip": {
|
||||
"Name": "tensorboard",
|
||||
"Version": "2.3.0"
|
||||
},
|
||||
"comments": "Installed in the training docker image"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
|
|
@ -493,15 +355,6 @@
|
|||
},
|
||||
"comments": "python-pillow. Implementation logic for anti-aliasing copied by Resize CPU kernel."
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "e7248b26a1ed53fa030c5c459f7ea095dfd276ac",
|
||||
"repositoryUrl": "https://gitlab.com/libeigen/eigen.git"
|
||||
}
|
||||
}
|
||||
}
|
||||
],
|
||||
"Version": 1
|
||||
|
|
|
|||
|
|
@ -1,159 +0,0 @@
|
|||
#!/usr/bin/env python3
|
||||
|
||||
import argparse
|
||||
import csv
|
||||
import json
|
||||
import os
|
||||
import re
|
||||
import subprocess
|
||||
import sys
|
||||
from dataclasses import dataclass
|
||||
from pathlib import PurePosixPath
|
||||
from urllib.parse import urlparse
|
||||
|
||||
import requests
|
||||
|
||||
|
||||
def parse_arguments():
|
||||
parser = argparse.ArgumentParser()
|
||||
parser.add_argument("--username", required=True, help="Github username")
|
||||
parser.add_argument("--token", required=True, help="Github access token")
|
||||
|
||||
return parser.parse_args()
|
||||
|
||||
|
||||
args = parse_arguments()
|
||||
|
||||
SCRIPT_DIR = os.path.dirname(os.path.realpath(__file__))
|
||||
REPO_DIR = os.path.normpath(os.path.join(SCRIPT_DIR, ".."))
|
||||
|
||||
package_name = None
|
||||
package_filename = None
|
||||
package_url = None
|
||||
|
||||
registrations = []
|
||||
|
||||
|
||||
@dataclass(frozen=True)
|
||||
class GitDep:
|
||||
commit: str
|
||||
url: str
|
||||
|
||||
|
||||
git_deps = {}
|
||||
|
||||
|
||||
def add_github_dep(name, parsed_url):
|
||||
segments = parsed_url.path.split("/")
|
||||
org_name = segments[1]
|
||||
repo_name = segments[2]
|
||||
if segments[3] != "archive":
|
||||
print("unrecognized github url path:" + parsed_url.path)
|
||||
return
|
||||
git_repo_url = f"https://github.com/{org_name}/{repo_name}.git"
|
||||
# For example, the path might be like '/myorg/myrepo/archive/5a5f8a5935762397aa68429b5493084ff970f774.zip'
|
||||
# The last segment, segments[4], is '5a5f8a5935762397aa68429b5493084ff970f774.zip'
|
||||
if len(segments) == 5 and re.match(r"[0-9a-f]{40}", PurePosixPath(segments[4]).stem):
|
||||
commit = PurePosixPath(segments[4]).stem
|
||||
dep = GitDep(commit, git_repo_url)
|
||||
if dep not in git_deps:
|
||||
git_deps[dep] = name
|
||||
else:
|
||||
# TODO: support urls like: https://github.com/onnx/onnx-tensorrt/archive/refs/tags/release/7.1.zip
|
||||
if len(segments) == 5:
|
||||
tag = PurePosixPath(segments[4]).stem
|
||||
if tag.endswith(".tar"):
|
||||
tag = PurePosixPath(tag).stem
|
||||
elif segments[4] == "refs" and segments[5] == "tags":
|
||||
tag = PurePosixPath(segments[6]).stem
|
||||
if tag.endswith(".tar"):
|
||||
tag = PurePosixPath(tag).stem
|
||||
else:
|
||||
print("unrecognized github url path:" + parsed_url.path)
|
||||
return
|
||||
# Make a REST call to convert to tag to a git commit
|
||||
url = f"https://api.github.com/repos/{org_name}/{repo_name}/git/refs/tags/{tag}"
|
||||
print(f"requesting {url} ...")
|
||||
res = requests.get(url, auth=(args.username, args.token))
|
||||
response_json = res.json()
|
||||
tag_object = response_json["object"]
|
||||
if tag_object["type"] == "commit":
|
||||
commit = tag_object["sha"]
|
||||
elif tag_object["type"] == "tag":
|
||||
res = requests.get(tag_object["url"], auth=(args.username, args.token))
|
||||
commit = res.json()["object"]["sha"]
|
||||
else:
|
||||
print("unrecognized github url path:" + parsed_url.path)
|
||||
return
|
||||
dep = GitDep(commit, git_repo_url)
|
||||
if dep not in git_deps:
|
||||
git_deps[dep] = name
|
||||
|
||||
|
||||
def normalize_path_separators(path):
|
||||
return path.replace(os.path.sep, "/")
|
||||
|
||||
|
||||
proc = subprocess.run(
|
||||
[
|
||||
"git",
|
||||
"submodule",
|
||||
"foreach",
|
||||
"--quiet",
|
||||
"'{}' '{}' $toplevel/$sm_path".format(
|
||||
normalize_path_separators(sys.executable),
|
||||
normalize_path_separators(os.path.join(SCRIPT_DIR, "print_submodule_info.py")),
|
||||
),
|
||||
],
|
||||
check=True,
|
||||
cwd=REPO_DIR,
|
||||
capture_output=True,
|
||||
text=True,
|
||||
)
|
||||
|
||||
|
||||
submodule_lines = proc.stdout.splitlines()
|
||||
for submodule_line in submodule_lines:
|
||||
(absolute_path, url, commit) = submodule_line.split(" ")
|
||||
git_deps[GitDep(commit, url)] = (
|
||||
f"git submodule at {normalize_path_separators(os.path.relpath(absolute_path, REPO_DIR))}"
|
||||
)
|
||||
|
||||
with open(os.path.join(SCRIPT_DIR, "..", "cmake", "deps.txt")) as f:
|
||||
depfile_reader = csv.reader(f, delimiter=";")
|
||||
for row in depfile_reader:
|
||||
if len(row) < 3:
|
||||
continue
|
||||
name = row[0]
|
||||
# Lines start with "#" are comments
|
||||
if name.startswith("#"):
|
||||
continue
|
||||
url = row[1]
|
||||
parsed_url = urlparse(url)
|
||||
# TODO: add support for gitlab
|
||||
if parsed_url.hostname == "github.com":
|
||||
add_github_dep(name, parsed_url)
|
||||
else:
|
||||
print("unrecognized url:" + url)
|
||||
|
||||
for git_dep, comment in git_deps.items():
|
||||
registration = {
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": git_dep.commit,
|
||||
"repositoryUrl": git_dep.url,
|
||||
},
|
||||
"comments": comment,
|
||||
}
|
||||
}
|
||||
registrations.append(registration)
|
||||
|
||||
cgmanifest = {
|
||||
"$schema": "https://json.schemastore.org/component-detection-manifest.json",
|
||||
"Version": 1,
|
||||
"Registrations": registrations,
|
||||
}
|
||||
|
||||
with open(os.path.join(SCRIPT_DIR, "generated", "cgmanifest.json"), mode="w") as generated_cgmanifest_file:
|
||||
print(json.dumps(cgmanifest, indent=2), file=generated_cgmanifest_file)
|
||||
|
|
@ -1,2 +0,0 @@
|
|||
`cgmanifests/generated/cgmanifest.json` is a generated file. Please do not edit it directly.
|
||||
More info [here](../README.md).
|
||||
|
|
@ -1,356 +0,0 @@
|
|||
{
|
||||
"$schema": "https://json.schemastore.org/component-detection-manifest.json",
|
||||
"Version": 1,
|
||||
"Registrations": [
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "d52c46520124845b1e0e0525f2759299d840143f",
|
||||
"repositoryUrl": "https://github.com/emscripten-core/emsdk.git"
|
||||
},
|
||||
"comments": "git submodule at cmake/external/emsdk"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "7a2ed51a6b682a83e345ff49fc4cfd7ca47550db",
|
||||
"repositoryUrl": "https://github.com/google/libprotobuf-mutator.git"
|
||||
},
|
||||
"comments": "git submodule at cmake/external/libprotobuf-mutator"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "b8baa8446686496da4cc8fda09f2b6fe65c2a02c",
|
||||
"repositoryUrl": "https://github.com/onnx/onnx.git"
|
||||
},
|
||||
"comments": "git submodule at cmake/external/onnx"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "4447c7562e3bc702ade25105912dce503f0c4010",
|
||||
"repositoryUrl": "https://github.com/abseil/abseil-cpp.git"
|
||||
},
|
||||
"comments": "abseil_cpp"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "dbb0094fd0cb936469e35320bf37e866ef7a1da4",
|
||||
"repositoryUrl": "https://github.com/apple/coremltools.git"
|
||||
},
|
||||
"comments": "coremltools"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "3c73d91c0b04e2b59462f0a741be8c07024c1bc0",
|
||||
"repositoryUrl": "https://github.com/jarro2783/cxxopts.git"
|
||||
},
|
||||
"comments": "cxxopts"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "6e921e1b1d21e84a5c82416ba7ecd98e33a436d0",
|
||||
"repositoryUrl": "https://github.com/HowardHinnant/date.git"
|
||||
},
|
||||
"comments": "date"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "277508879878e0a5b5b43599b1bea11f66eb3c6c",
|
||||
"repositoryUrl": "https://github.com/dmlc/dlpack.git"
|
||||
},
|
||||
"comments": "dlpack"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "0100f6a5779831fa7a651e4b67ef389a8752bd9b",
|
||||
"repositoryUrl": "https://github.com/google/flatbuffers.git"
|
||||
},
|
||||
"comments": "flatbuffers"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "0a92994d729ff76a58f692d3028ca1b64b145d91",
|
||||
"repositoryUrl": "https://github.com/Maratyszcza/FP16.git"
|
||||
},
|
||||
"comments": "fp16"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "63058eff77e11aa15bf531df5dd34395ec3017c8",
|
||||
"repositoryUrl": "https://github.com/Maratyszcza/FXdiv.git"
|
||||
},
|
||||
"comments": "fxdiv"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "a6ad7fbbdc2e14fab82bb8a6d27760d700198cbf",
|
||||
"repositoryUrl": "https://github.com/google/benchmark.git"
|
||||
},
|
||||
"comments": "google_benchmark"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "e39786088138f2749d64e9e90e0f9902daa77c40",
|
||||
"repositoryUrl": "https://github.com/google/googletest.git"
|
||||
},
|
||||
"comments": "googletest"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "fe98e0b93565382648129271381c14d6205255e3",
|
||||
"repositoryUrl": "https://github.com/google/XNNPACK.git"
|
||||
},
|
||||
"comments": "googlexnnpack"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "4f8fba14066156b73f1189a2b8bd568bde5284c5",
|
||||
"repositoryUrl": "https://github.com/nlohmann/json.git"
|
||||
},
|
||||
"comments": "json"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "a3534567187d2edc428efd3f13466ff75fe5805c",
|
||||
"repositoryUrl": "https://github.com/microsoft/GSL.git"
|
||||
},
|
||||
"comments": "microsoft_gsl"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "55f373e164d3f092dd6c7a56e3de6f90c4c6f3dc",
|
||||
"repositoryUrl": "https://github.com/microsoft/wil.git"
|
||||
},
|
||||
"comments": "microsoft_wil"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "3e313478d91c04ac5821743688ce55fc27432c4f",
|
||||
"repositoryUrl": "https://github.com/microsoft/mimalloc.git"
|
||||
},
|
||||
"comments": "mimalloc"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "0a0b5fb001ce0233ae3a6f99d849c0649e5a7361",
|
||||
"repositoryUrl": "https://github.com/boostorg/mp11.git"
|
||||
},
|
||||
"comments": "mp11"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "9c69a24bc2e20c8a511a4e6b06fd49639ec5300a",
|
||||
"repositoryUrl": "https://github.com/onnx/onnx-tensorrt.git"
|
||||
},
|
||||
"comments": "onnx_tensorrt"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "f0dc78d7e6e331b8c6bb2d5283e06aa26883ca7c",
|
||||
"repositoryUrl": "https://github.com/protocolbuffers/protobuf.git"
|
||||
},
|
||||
"comments": "protobuf"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "072586a71b55b7f8c584153d223e95687148a900",
|
||||
"repositoryUrl": "https://github.com/Maratyszcza/psimd.git"
|
||||
},
|
||||
"comments": "psimd"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "4e80ca24521aa0fb3a746f9ea9c3eaa20e9afbb0",
|
||||
"repositoryUrl": "https://github.com/google/pthreadpool.git"
|
||||
},
|
||||
"comments": "pthreadpool"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "941f45bcb51457884fa1afd6e24a67377d70f75c",
|
||||
"repositoryUrl": "https://github.com/pybind/pybind11.git"
|
||||
},
|
||||
"comments": "pybind11"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "8a1772a0c5c447df2d18edf33ec4603a8c9c04a6",
|
||||
"repositoryUrl": "https://github.com/pytorch/cpuinfo.git"
|
||||
},
|
||||
"comments": "pytorch_cpuinfo"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "6dcd83d60f7944926bfd308cc13979fc53dd69ca",
|
||||
"repositoryUrl": "https://github.com/google/re2.git"
|
||||
},
|
||||
"comments": "re2"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "4cafc9196c4da9c817992b20f5253ef967685bf8",
|
||||
"repositoryUrl": "https://github.com/dcleblanc/SafeInt.git"
|
||||
},
|
||||
"comments": "safeint"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "373eb09e4c5d2b3cc2493f0949dc4be6b6a45e81",
|
||||
"repositoryUrl": "https://github.com/tensorflow/tensorboard.git"
|
||||
},
|
||||
"comments": "tensorboard"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "f7b19de32c5d1f3cedfc735c2849f12b537522ee",
|
||||
"repositoryUrl": "https://github.com/NVIDIA/cutlass.git"
|
||||
},
|
||||
"comments": "cutlass"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "72c943dea2b9240cd09efde15191e144bc7c7d38",
|
||||
"repositoryUrl": "https://github.com/protocolbuffers/utf8_range.git"
|
||||
},
|
||||
"comments": "utf8_range"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "f3f6caa6e8adb420e005ec41c6fefc8d75affb6e",
|
||||
"repositoryUrl": "https://github.com/microsoft/onnxruntime-extensions.git"
|
||||
},
|
||||
"comments": "extensions"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "204da9c522cebec5220bba52cd3542ebcaf99e7a",
|
||||
"repositoryUrl": "https://github.com/ROCmSoftwarePlatform/composable_kernel.git"
|
||||
},
|
||||
"comments": "composable_kernel"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "de28d93dfa9ebf3e473127c1c657e1920a5345ee",
|
||||
"repositoryUrl": "https://github.com/microsoft/DirectX-Headers.git"
|
||||
},
|
||||
"comments": "directx_headers"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "de355c7094af70467f2b264f531ab5c5f4401c42",
|
||||
"repositoryUrl": "https://github.com/NVIDIA/cudnn-frontend.git"
|
||||
},
|
||||
"comments": "cudnn_frontend"
|
||||
}
|
||||
},
|
||||
{
|
||||
"component": {
|
||||
"type": "git",
|
||||
"git": {
|
||||
"commitHash": "b9b4a37041dec3dd62ac92014a6cc1aece48d9f3",
|
||||
"repositoryUrl": "https://github.com/google/dawn.git"
|
||||
},
|
||||
"comments": "dawn"
|
||||
}
|
||||
}
|
||||
]
|
||||
}
|
||||
|
|
@ -36,8 +36,8 @@ microsoft_wil;https://github.com/microsoft/wil/archive/refs/tags/v1.0.230629.1.z
|
|||
mimalloc;https://github.com/microsoft/mimalloc/archive/refs/tags/v2.1.1.zip;d5ee7d34223d0567892db5179849939c8769dc41
|
||||
mp11;https://github.com/boostorg/mp11/archive/refs/tags/boost-1.82.0.zip;9bc9e01dffb64d9e0773b2e44d2f22c51aace063
|
||||
onnx;https://github.com/onnx/onnx/archive/refs/tags/v1.17.0.zip;13a60ac5217c104139ce0fd024f48628e7bcf5bc
|
||||
# Use the latest commit of 10.7-GA
|
||||
onnx_tensorrt;https://github.com/onnx/onnx-tensorrt/archive/9c69a24bc2e20c8a511a4e6b06fd49639ec5300a.zip;ff1fe9af78eb129b4a4cdcb7450b7390b4436dd3
|
||||
# Use the latest commit of 10.8-GA
|
||||
onnx_tensorrt;https://github.com/onnx/onnx-tensorrt/archive/118ed0aea197fa9a7d3ea66180a1d5ddb9deecc3.zip;b78aed3728ad4daf6dc47ea10c1d243dee1d95b1
|
||||
protobuf;https://github.com/protocolbuffers/protobuf/archive/refs/tags/v21.12.zip;7cf2733949036c7d52fda017badcab093fe73bfa
|
||||
protoc_win64;https://github.com/protocolbuffers/protobuf/releases/download/v21.12/protoc-21.12-win64.zip;b4521f7ada5b260380f94c4bd7f1b7684c76969a
|
||||
protoc_win32;https://github.com/protocolbuffers/protobuf/releases/download/v21.12/protoc-21.12-win32.zip;3688010318192c46ce73213cdfb6b3e5656da874
|
||||
|
|
|
|||
|
|
@ -600,6 +600,7 @@ endif()
|
|||
|
||||
if(onnxruntime_ENABLE_DLPACK)
|
||||
message(STATUS "dlpack is enabled.")
|
||||
|
||||
onnxruntime_fetchcontent_declare(
|
||||
dlpack
|
||||
URL ${DEP_URL_dlpack}
|
||||
|
|
@ -607,9 +608,7 @@ if(onnxruntime_ENABLE_DLPACK)
|
|||
EXCLUDE_FROM_ALL
|
||||
FIND_PACKAGE_ARGS NAMES dlpack
|
||||
)
|
||||
# We can't use onnxruntime_fetchcontent_makeavailable since some part of the the dlpack code is Linux only.
|
||||
# For example, dlpackcpp.h uses posix_memalign.
|
||||
FetchContent_Populate(dlpack)
|
||||
onnxruntime_fetchcontent_makeavailable(dlpack)
|
||||
endif()
|
||||
|
||||
if(onnxruntime_ENABLE_TRAINING OR (onnxruntime_ENABLE_TRAINING_APIS AND onnxruntime_BUILD_UNIT_TESTS))
|
||||
|
|
|
|||
|
|
@ -77,8 +77,7 @@ endif()
|
|||
if (onnxruntime_ENABLE_TRAINING_OPS)
|
||||
target_include_directories(onnxruntime_framework PRIVATE ${ORTTRAINING_ROOT})
|
||||
if (onnxruntime_ENABLE_TRAINING_TORCH_INTEROP OR onnxruntime_ENABLE_TRITON)
|
||||
onnxruntime_add_include_to_target(onnxruntime_framework Python::Module)
|
||||
target_include_directories(onnxruntime_framework PRIVATE ${dlpack_SOURCE_DIR}/include)
|
||||
onnxruntime_add_include_to_target(onnxruntime_framework Python::Module dlpack::dlpack)
|
||||
endif()
|
||||
endif()
|
||||
if (onnxruntime_USE_MPI)
|
||||
|
|
@ -86,9 +85,7 @@ if (onnxruntime_USE_MPI)
|
|||
endif()
|
||||
|
||||
if (onnxruntime_ENABLE_ATEN)
|
||||
# DLPack is a header-only dependency
|
||||
set(DLPACK_INCLUDE_DIR ${dlpack_SOURCE_DIR}/include)
|
||||
target_include_directories(onnxruntime_framework PRIVATE ${DLPACK_INCLUDE_DIR})
|
||||
onnxruntime_add_include_to_target(onnxruntime_framework dlpack::dlpack)
|
||||
endif()
|
||||
onnxruntime_add_include_to_target(onnxruntime_framework onnxruntime_common onnx onnx_proto ${PROTOBUF_LIB} flatbuffers::flatbuffers safeint_interface Boost::mp11 nlohmann_json::nlohmann_json)
|
||||
|
||||
|
|
|
|||
|
|
@ -196,8 +196,7 @@ endif()
|
|||
if (onnxruntime_ENABLE_DLPACK)
|
||||
target_compile_definitions(onnxruntime_providers PRIVATE ENABLE_DLPACK)
|
||||
# DLPack is a header-only dependency
|
||||
set(DLPACK_INCLUDE_DIR ${dlpack_SOURCE_DIR}/include)
|
||||
target_include_directories(onnxruntime_providers PRIVATE ${DLPACK_INCLUDE_DIR})
|
||||
onnxruntime_add_include_to_target(onnxruntime_providers dlpack::dlpack)
|
||||
endif()
|
||||
|
||||
if (onnxruntime_ENABLE_TRAINING)
|
||||
|
|
|
|||
|
|
@ -131,7 +131,7 @@ if (onnxruntime_ENABLE_ATEN)
|
|||
endif()
|
||||
|
||||
if (onnxruntime_ENABLE_DLPACK)
|
||||
target_include_directories(onnxruntime_pybind11_state PRIVATE ${dlpack_SOURCE_DIR}/include)
|
||||
target_link_libraries(onnxruntime_pybind11_state PRIVATE dlpack::dlpack)
|
||||
endif()
|
||||
|
||||
if (onnxruntime_ENABLE_TRAINING)
|
||||
|
|
|
|||
|
|
@ -250,6 +250,12 @@ static const char* const kOrtSessionOptionsOptimizedModelExternalInitializersFil
|
|||
static const char* const kOrtSessionOptionsOptimizedModelExternalInitializersMinSizeInBytes =
|
||||
"session.optimized_model_external_initializers_min_size_in_bytes";
|
||||
|
||||
// When loading model from memory buffer and the model has external initializers
|
||||
// Use this config to set the external data file folder path
|
||||
// All external data files should be in the same folder
|
||||
static const char* const kOrtSessionOptionsModelExternalInitializersFileFolderPath =
|
||||
"session.model_external_initializers_file_folder_path";
|
||||
|
||||
// Use this config when saving pre-packed constant initializers to an external data file.
|
||||
// This allows you to memory map pre-packed initializers on model load and leave it to
|
||||
// to the OS the amount of memory consumed by the pre-packed initializers. Otherwise,
|
||||
|
|
|
|||
|
|
@ -46,11 +46,6 @@ export const createConvTranspose2DProgramInfo = (
|
|||
const inputChannelsPerGroup = wShape[2] / group;
|
||||
const outputChannelsPerGroup = wShape[3];
|
||||
const aComponents = isChannelsLast ? getMaxComponents(inputChannelsPerGroup) : 1;
|
||||
const packInputAs4 = isChannelsLast && outputChannelsPerGroup === 1;
|
||||
const inputChannelsPerGroupInt = packInputAs4
|
||||
? Math.floor(inputChannelsPerGroup / 4) * 4
|
||||
: Math.floor(inputChannelsPerGroup / aComponents) * aComponents;
|
||||
const inputChannelsRemainder = inputChannelsPerGroup - inputChannelsPerGroupInt;
|
||||
const components = isChannelsLast ? getMaxComponents(outputChannelsPerGroup) : 1;
|
||||
const bComponents = isChannelsLast ? (outputChannelsPerGroup === 1 ? aComponents : components) : 1;
|
||||
const outputSize = ShapeUtil.size(outputShape) / components;
|
||||
|
|
@ -83,7 +78,7 @@ export const createConvTranspose2DProgramInfo = (
|
|||
{ type: DataType.uint32, data: dilations },
|
||||
{ type: DataType.uint32, data: effectiveFilterDims },
|
||||
{ type: DataType.int32, data: pads },
|
||||
{ type: DataType.uint32, data: inputChannelsPerGroupInt },
|
||||
{ type: DataType.uint32, data: inputChannelsPerGroup },
|
||||
{ type: DataType.uint32, data: outputChannelsPerGroup },
|
||||
...createTensorShapeVariables(inputs[0].dims, inputs[1].dims),
|
||||
];
|
||||
|
|
@ -119,40 +114,16 @@ export const createConvTranspose2DProgramInfo = (
|
|||
|
||||
const calculateResult = (): string => {
|
||||
let calcStr = '';
|
||||
if (packInputAs4) {
|
||||
if (aComponents === 4) {
|
||||
calcStr += `
|
||||
let xValue = ${dy.getByOffset('x_offset')};
|
||||
let wValue = ${w.getByOffset('w_offset')};
|
||||
dotProd = dotProd + dot(xValue, wValue);
|
||||
x_offset += 1u;
|
||||
w_offset += 1u;`;
|
||||
} else if (aComponents === 2) {
|
||||
calcStr += `
|
||||
dotProd = dotProd + dot(vec4<${dataType}>(${dy.getByOffset('x_offset')}, ${dy.getByOffset('x_offset + 1u')}), vec4<${dataType}>(${w.getByOffset('w_offset')}, ${w.getByOffset('w_offset + 1u')}));
|
||||
x_offset += 2u;
|
||||
w_offset += 2u;`;
|
||||
} else if (aComponents === 1) {
|
||||
calcStr += `
|
||||
dotProd = dotProd + dot(vec4<${dataType}>(${dy.getByOffset('x_offset')}, ${dy.getByOffset('x_offset + 1u')}, ${dy.getByOffset('x_offset + 2u')}, ${dy.getByOffset('x_offset + 3u')}), vec4<${dataType}>(${w.getByOffset('w_offset')}, ${w.getByOffset('w_offset + 1u')}, ${w.getByOffset('w_offset + 2u')}, ${w.getByOffset('w_offset + 3u')}));
|
||||
x_offset += 4u;
|
||||
w_offset += 4u;`;
|
||||
}
|
||||
} else {
|
||||
if (aComponents === 1) {
|
||||
calcStr += `
|
||||
let xValue = ${
|
||||
isChannelsLast
|
||||
? dy.getByOffset(
|
||||
`${dy.indicesToOffset(`${dy.type.indices}(batch, idyR, idyC, inputChannel)`)} / ${aComponents}`,
|
||||
)
|
||||
: dy.get('batch', 'inputChannel', 'idyR', 'idyC')
|
||||
};
|
||||
`;
|
||||
if (aComponents === 1) {
|
||||
let w_offset = ${w.indicesToOffset(`${w.type.indices}(u32(wRPerm), u32(wCPerm), inputChannel, wOutChannel)`)};
|
||||
let wValue = ${w.getByOffset(`w_offset / ${bComponents}`)};
|
||||
dotProd = dotProd + xValue * wValue;`;
|
||||
} else {
|
||||
if (outputChannelsPerGroup === 1) {
|
||||
calcStr += `
|
||||
let w_offset = ${w.indicesToOffset(`${w.type.indices}(u32(wRPerm), u32(wCPerm), inputChannel, wOutChannel)`)};
|
||||
let wValue = ${w.getByOffset(`w_offset / ${bComponents}`)};
|
||||
dotProd = dotProd + xValue * wValue;`;
|
||||
let wValue = ${w.getByOffset(`${w.indicesToOffset(`${w.type.indices}(u32(wRPerm), u32(wCPerm), inputChannel, wOutChannel)`)} / ${bComponents}`)};
|
||||
dotProd = dotProd + dot(xValue, wValue);`;
|
||||
} else {
|
||||
for (let c = 0; c < aComponents; c++) {
|
||||
calcStr += `
|
||||
|
|
@ -163,32 +134,6 @@ export const createConvTranspose2DProgramInfo = (
|
|||
}
|
||||
return calcStr;
|
||||
};
|
||||
const calculateRemainder = (): string => {
|
||||
if (inputChannelsRemainder === 0) {
|
||||
return '';
|
||||
}
|
||||
if (!packInputAs4) {
|
||||
throw new Error(`packInputAs4 ${packInputAs4} is not true.`);
|
||||
}
|
||||
let calcStr = '';
|
||||
if (aComponents === 1) {
|
||||
calcStr += 'dotProd = dotProd';
|
||||
for (let i = 0; i < inputChannelsRemainder; i++) {
|
||||
calcStr += `
|
||||
+ ${dy.getByOffset(`x_offset + ${i}`)} * ${w.getByOffset(`w_offset + ${i}`)}`;
|
||||
}
|
||||
calcStr += ';';
|
||||
} else if (aComponents === 2) {
|
||||
if (inputChannelsRemainder !== 2) {
|
||||
throw new Error(`Invalid inputChannelsRemainder ${inputChannelsRemainder}.`);
|
||||
}
|
||||
calcStr += `
|
||||
let xValue = ${dy.getByOffset('x_offset')};
|
||||
let wValue = ${w.getByOffset('w_offset')};
|
||||
dotProd = dotProd + dot(xValue, wValue);`;
|
||||
}
|
||||
return calcStr;
|
||||
};
|
||||
const codeSnippet = `
|
||||
let outputIndices = ${output.offsetToIndices(`global_idx * ${components}`)};
|
||||
let batch = ${output.indicesGet('outputIndices', 0)};
|
||||
|
|
@ -224,6 +169,7 @@ export const createConvTranspose2DProgramInfo = (
|
|||
// Minimum wC >= 0 that satisfies (dyCCorner + wC) % (uniforms.strides.y) == 0
|
||||
wC = u32(((dyCCorner + i32(uniforms.strides.y) - 1) / i32(uniforms.strides.y)) * i32(uniforms.strides.y) - dyCCorner);
|
||||
}
|
||||
|
||||
for (; wC < uniforms.effective_filter_dims.y; wC = wC + 1) {
|
||||
if (wC % uniforms.dilations.y != 0) {
|
||||
continue;
|
||||
|
|
@ -236,19 +182,17 @@ export const createConvTranspose2DProgramInfo = (
|
|||
}
|
||||
let idyC: u32 = u32(dyC);
|
||||
var inputChannel = groupId * uniforms.input_channels_per_group;
|
||||
${
|
||||
packInputAs4
|
||||
? `
|
||||
var x_offset = ${dy.indicesToOffset(`${dy.type.indices}(batch, idyR, idyC, inputChannel)`)} / ${aComponents};
|
||||
var w_offset = ${w.indicesToOffset(`${w.type.indices}(wRPerm, wCPerm, inputChannel, wOutChannel)`)} / ${bComponents};
|
||||
`
|
||||
: ''
|
||||
}
|
||||
for (var d2: u32 = 0; d2 < uniforms.input_channels_per_group; d2 = d2 + ${packInputAs4 ? 4 : aComponents}) {
|
||||
for (var d2: u32 = 0; d2 < uniforms.input_channels_per_group; d2 = d2 + ${aComponents}) {
|
||||
let xValue = ${
|
||||
isChannelsLast
|
||||
? dy.getByOffset(
|
||||
`${dy.indicesToOffset(`${dy.type.indices}(batch, idyR, idyC, inputChannel)`)} / ${aComponents}`,
|
||||
)
|
||||
: dy.get('batch', 'inputChannel', 'idyR', 'idyC')
|
||||
};
|
||||
${calculateResult()}
|
||||
inputChannel = inputChannel + ${packInputAs4 ? 4 : aComponents};
|
||||
inputChannel = inputChannel + ${aComponents};
|
||||
}
|
||||
${calculateRemainder()}
|
||||
wC = wC + uniforms.strides.y - 1;
|
||||
}
|
||||
wR = wR + uniforms.strides[0] - 1;
|
||||
|
|
@ -267,7 +211,7 @@ export const createConvTranspose2DProgramInfo = (
|
|||
return {
|
||||
name: 'ConvTranspose2D',
|
||||
shaderCache: {
|
||||
hint: `${attributes.cacheKey};${aComponents}${bComponents}${components}${outputChannelsPerGroup === 1}${inputChannelsRemainder}`,
|
||||
hint: `${attributes.cacheKey};${aComponents}${bComponents}${components}${outputChannelsPerGroup === 1}`,
|
||||
inputDependencies,
|
||||
},
|
||||
getRunData: () => ({
|
||||
|
|
|
|||
|
|
@ -291,8 +291,10 @@ const convTranspose1d = (context: ComputeContext, attributes: ConvTransposeAttri
|
|||
strides = [1].concat(strides);
|
||||
dilations = [1].concat(dilations);
|
||||
kernelShape = [1].concat(kernelShape);
|
||||
let outputPadding = attributes.outputPadding;
|
||||
outputPadding = [0].concat(outputPadding);
|
||||
const adjustedAttributes = getAdjustedConvTransposeAttributes(
|
||||
{ ...attributes, pads, strides, dilations, kernelShape },
|
||||
{ ...attributes, pads, strides, dilations, kernelShape, outputPadding },
|
||||
inputs,
|
||||
);
|
||||
|
||||
|
|
|
|||
|
|
@ -157,6 +157,16 @@ const validateInputs = (
|
|||
}
|
||||
};
|
||||
|
||||
const getSafeIntegerDivision = (a: string, b: string, c: string, dType: string): string => `
|
||||
// The whole part and the fractional part are calculated separately due to inaccuracy of floating
|
||||
// point division. As an example, f32(21) / f32(7) may evaluate to 2.99... instead of 3, causing an
|
||||
// offset-by-one error later in floor().
|
||||
let big = (${a}) * (${b});
|
||||
let whole = ${dType}(big / (${c}));
|
||||
let fract = ${dType}(big % (${c})) / ${dType}(${c});
|
||||
return whole + fract;
|
||||
`;
|
||||
|
||||
const getOriginalCoordinateFromResizedCoordinate = (
|
||||
coordinateTransferMode: CoordinateTransformMode,
|
||||
dType: string,
|
||||
|
|
@ -166,7 +176,13 @@ const getOriginalCoordinateFromResizedCoordinate = (
|
|||
(() => {
|
||||
switch (coordinateTransferMode) {
|
||||
case 'asymmetric':
|
||||
return `return ${dType}(xResized) / ${dType}(xScale);`;
|
||||
return `
|
||||
if (xScale < 1.0 || floor(xScale) != xScale) {
|
||||
return ${dType}(xResized) / ${dType}(xScale);
|
||||
} else {
|
||||
${getSafeIntegerDivision('xResized', 'lengthOriginal', 'lengthResized', dType)}
|
||||
}
|
||||
`;
|
||||
case 'pytorch_half_pixel':
|
||||
return `if (lengthResized > 1) {
|
||||
return (${dType}(xResized) + 0.5) / ${dType}(xScale) - 0.5;
|
||||
|
|
@ -179,13 +195,7 @@ const getOriginalCoordinateFromResizedCoordinate = (
|
|||
return `if (lengthResized == 1) {
|
||||
return 0.0;
|
||||
} else {
|
||||
// The whole part and the fractional part are calculated separately due to inaccuracy of floating
|
||||
// point division. As an example, f32(21) / f32(7) may evaluate to 2.99... instead of 3, causing an
|
||||
// offset-by-one error later in floor().
|
||||
let whole = ${dType}(xResized * (lengthOriginal - 1) / (lengthResized - 1));
|
||||
let fract =
|
||||
${dType}(xResized * (lengthOriginal - 1) % (lengthResized - 1)) / ${dType}(lengthResized - 1);
|
||||
return whole + fract;
|
||||
${getSafeIntegerDivision('xResized', 'lengthOriginal - 1', 'lengthResized - 1', dType)}
|
||||
}`;
|
||||
case 'tf_crop_and_resize':
|
||||
return `if (lengthResized > 1) {
|
||||
|
|
@ -375,7 +385,7 @@ const calculateInputIndicesFromOutputIndices = (
|
|||
input_index = u32(original_idx);
|
||||
}
|
||||
}
|
||||
${input.indicesSet('input_indices', 'i', ' input_index')}
|
||||
${input.indicesSet('input_indices', 'i', 'input_index')}
|
||||
}
|
||||
return input_indices;
|
||||
}`;
|
||||
|
|
@ -758,9 +768,11 @@ const createResizeProgramInfo = (
|
|||
return {
|
||||
name: 'Resize',
|
||||
shaderCache: {
|
||||
hint: `${attributes.cacheKey}|${opsetVersion}|${scales.length > 0 ? scales : ''}|${
|
||||
sizes.length > 0 ? sizes : ''
|
||||
}|${roi.length > 0 ? roi : ''}|${noScale}|${inputShape}`,
|
||||
hint: `${attributes.cacheKey}|${opsetVersion}|${
|
||||
scales.length > 0 ? (attributes.mode === 'cubic' ? scales : scales.length) : ''
|
||||
}|${sizes.length > 0 ? sizes : ''}|${roi.length > 0 ? roi : ''}|${noScale}|${
|
||||
attributes.mode === 'nearest' ? inputShape.length : inputShape
|
||||
}`,
|
||||
inputDependencies: ['rank'],
|
||||
},
|
||||
getShaderSource,
|
||||
|
|
|
|||
|
|
@ -865,5 +865,82 @@
|
|||
]
|
||||
}
|
||||
]
|
||||
},
|
||||
{
|
||||
"name": "ConvTranspose1D with output_padding",
|
||||
"operator": "ConvTranspose",
|
||||
"opset": {
|
||||
"domain": "",
|
||||
"version": 17
|
||||
},
|
||||
"attributes": [
|
||||
{
|
||||
"name": "dilations",
|
||||
"data": [1],
|
||||
"type": "ints"
|
||||
},
|
||||
{
|
||||
"name": "group",
|
||||
"data": 4,
|
||||
"type": "int"
|
||||
},
|
||||
{
|
||||
"name": "kernel_shape",
|
||||
"data": [3],
|
||||
"type": "ints"
|
||||
},
|
||||
{
|
||||
"name": "output_padding",
|
||||
"data": [1],
|
||||
"type": "ints"
|
||||
},
|
||||
{
|
||||
"name": "pads",
|
||||
"data": [1, 1],
|
||||
"type": "ints"
|
||||
},
|
||||
{
|
||||
"name": "strides",
|
||||
"data": [2],
|
||||
"type": "ints"
|
||||
}
|
||||
],
|
||||
"cases": [
|
||||
{
|
||||
"name": "ConvTranspose1D with output_padding - 1",
|
||||
"inputs": [
|
||||
{
|
||||
"dims": [1, 4, 8],
|
||||
"type": "float32",
|
||||
"data": [
|
||||
1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16, 17, 18, 19, 20, 21, 22, 23, 24, 25, 26, 27, 28, 29,
|
||||
30, 31, 32
|
||||
]
|
||||
},
|
||||
{
|
||||
"dims": [4, 1, 3],
|
||||
"type": "float32",
|
||||
"data": [1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12]
|
||||
},
|
||||
{
|
||||
"dims": [4],
|
||||
"type": "float32",
|
||||
"data": [1, 2, 3, 4]
|
||||
}
|
||||
],
|
||||
"outputs": [
|
||||
{
|
||||
"name": "output",
|
||||
"dims": [1, 4, 16],
|
||||
"type": "float32",
|
||||
"data": [
|
||||
3, 6, 5, 10, 7, 14, 9, 18, 11, 22, 13, 26, 15, 30, 17, 25, 47, 96, 52, 106, 57, 116, 62, 126, 67, 136, 72,
|
||||
146, 77, 156, 82, 98, 139, 282, 147, 298, 155, 314, 163, 330, 171, 346, 179, 362, 187, 378, 195, 219, 279,
|
||||
564, 290, 586, 301, 608, 312, 630, 323, 652, 334, 674, 345, 696, 356, 388
|
||||
]
|
||||
}
|
||||
]
|
||||
}
|
||||
]
|
||||
}
|
||||
]
|
||||
|
|
|
|||
|
|
@ -1347,6 +1347,7 @@
|
|||
"concat_zero-sized.jsonc",
|
||||
"cast.jsonc",
|
||||
"conv.jsonc",
|
||||
"conv-transpose.jsonc",
|
||||
"conv1d.jsonc",
|
||||
"conv3dncdhw.jsonc",
|
||||
"cos.jsonc",
|
||||
|
|
|
|||
|
|
@ -8,7 +8,7 @@
|
|||
#include <gsl/gsl>
|
||||
#include "core/framework/allocator.h"
|
||||
#include "core/framework/ort_value.h"
|
||||
#include "contrib_ops/cpu/utils/debug_macros.h"
|
||||
#include "contrib_ops/cpu/utils/console_dumper.h"
|
||||
|
||||
namespace onnxruntime {
|
||||
|
||||
|
|
|
|||
|
|
@ -1,11 +1,10 @@
|
|||
// Copyright (c) Microsoft Corporation. All rights reserved.
|
||||
// Licensed under the MIT License.
|
||||
|
||||
|
||||
// cub.cuh includes device/dispatch_radix_sort.cuh which has assignment in conditional expressions
|
||||
#if defined(_MSC_VER)
|
||||
#pragma warning(push)
|
||||
#pragma warning(disable : 4706)
|
||||
#pragma warning(disable : 4706)
|
||||
#endif
|
||||
#include <cub/cub.cuh>
|
||||
#if defined(_MSC_VER)
|
||||
|
|
@ -323,8 +322,8 @@ __device__ void BeamHypotheses::Output(
|
|||
int top_k,
|
||||
int max_length,
|
||||
int pad_token_id,
|
||||
int32_t* sequences, // buffer of shape (num_return_sequences, max_length)
|
||||
T* sequences_scores) // buffer of shape (num_return_sequences) or empty
|
||||
int32_t* sequences, // buffer of shape (num_return_sequences, max_length)
|
||||
T* sequences_scores) // buffer of shape (num_return_sequences) or empty
|
||||
{
|
||||
// Copy the top_k beams into the sequences
|
||||
for (int index = 0; index < top_k; index++) {
|
||||
|
|
@ -343,6 +342,7 @@ __device__ void BeamHypotheses::Output(
|
|||
}
|
||||
}
|
||||
|
||||
template <bool early_stopping>
|
||||
__global__ void BeamSearchScorer_Process(BeamScorerState& state_cpu,
|
||||
BeamScorerState& state,
|
||||
const int32_t* sequences_buffer,
|
||||
|
|
@ -358,24 +358,26 @@ __global__ void BeamSearchScorer_Process(BeamScorerState& state_cpu,
|
|||
// Sequences shape is (batch_size * num_beams, total_sequence_length)
|
||||
// It contains word ID of whole sequence generated so far.
|
||||
// It is different from subgraph input_ids, which only need one word when past state is not empty.
|
||||
|
||||
int batch = threadIdx.x;
|
||||
int batch_start = batch * state.num_beams_;
|
||||
const int batch = threadIdx.x;
|
||||
const int num_beams = state.num_beams_;
|
||||
const int batch_start = batch * num_beams;
|
||||
|
||||
cuda::BeamHypotheses& beam_hyp = beam_hyps_[batch];
|
||||
|
||||
if (!beam_hyp.done_) {
|
||||
// Next tokens for this sentence.
|
||||
size_t beam_idx = 0;
|
||||
size_t top_k = 2 * state.num_beams_;
|
||||
size_t top_k = 2 * num_beams;
|
||||
for (size_t j = 0; j < top_k; j++) {
|
||||
int32_t next_token = next_tokens[batch * top_k + j];
|
||||
float next_score = next_scores[batch * top_k + j];
|
||||
int32_t next_index = next_indices[batch * top_k + j];
|
||||
|
||||
int batch_beam_idx = batch_start + next_index;
|
||||
|
||||
// Add to generated hypotheses if end of sentence.
|
||||
if ((state.eos_token_id_ >= 0) && (next_token == state.eos_token_id_)) {
|
||||
bool is_beam_token_worse_than_top_num_beams = (j >= state.num_beams_);
|
||||
bool is_beam_token_worse_than_top_num_beams = (j >= num_beams);
|
||||
if (is_beam_token_worse_than_top_num_beams) {
|
||||
continue;
|
||||
}
|
||||
|
|
@ -396,21 +398,27 @@ __global__ void BeamSearchScorer_Process(BeamScorerState& state_cpu,
|
|||
}
|
||||
|
||||
// Once the beam for next step is full, don't add more tokens to it.
|
||||
if (beam_idx == state.num_beams_)
|
||||
if (beam_idx == num_beams)
|
||||
break;
|
||||
}
|
||||
|
||||
// Check if we are done so that we can save a pad step if all(done)
|
||||
if (beam_hyp.beams_used_ == state.num_beams_) {
|
||||
if (state.early_stopping_ || !beam_hyp.CanImprove(*std::max_element(next_scores + batch_start, next_scores + batch_start + top_k), sequence_length)) {
|
||||
beam_hyp.done_ = true;
|
||||
if (atomicAdd(&state.not_done_count_, -1) == 1)
|
||||
state_cpu.not_done_count_ = 0; // Update the CPU side
|
||||
if (beam_hyp.beams_used_ == num_beams) {
|
||||
if constexpr (!early_stopping) {
|
||||
float best_sum_logprobs = *std::max_element(next_scores + batch_start, next_scores + batch_start + top_k);
|
||||
if (beam_hyp.CanImprove(best_sum_logprobs, sequence_length)) {
|
||||
return;
|
||||
}
|
||||
}
|
||||
|
||||
beam_hyp.done_ = true;
|
||||
if (atomicAdd(&(state.not_done_count_), -1) == 1) {
|
||||
state_cpu.not_done_count_ = 0; // Update the CPU side
|
||||
}
|
||||
}
|
||||
} else {
|
||||
// Pad the batch.
|
||||
for (size_t beam_idx = 0; beam_idx < state.num_beams_; beam_idx++) {
|
||||
for (size_t beam_idx = 0; beam_idx < num_beams; beam_idx++) {
|
||||
next_beam_scores_[batch_start + beam_idx] = 0.0f;
|
||||
next_beam_tokens_[batch_start + beam_idx] = state.pad_token_id_;
|
||||
next_beam_indices_[batch_start + beam_idx] = 0;
|
||||
|
|
@ -418,6 +426,31 @@ __global__ void BeamSearchScorer_Process(BeamScorerState& state_cpu,
|
|||
}
|
||||
}
|
||||
|
||||
#ifdef DEBUG_GENERATION
|
||||
__global__ void DumpBeamScorerState(const BeamScorerState& state) {
|
||||
state.Print(false);
|
||||
}
|
||||
|
||||
void DumpBeamScorerStates(const BeamScorerState& state_cpu, const BeamScorerState& state, cudaStream_t stream) {
|
||||
state_cpu.Print(true);
|
||||
DumpBeamScorerState<<<1, 1, 0, stream>>>(state);
|
||||
cudaDeviceSynchronize();
|
||||
}
|
||||
|
||||
__global__ void DumpBeamSearchScorer(const BeamHypotheses& beam_hyp) {
|
||||
beam_hyp.Print();
|
||||
}
|
||||
|
||||
void DumpBeamHypotheses(gsl::span<BeamHypotheses> beam_hyps, cudaStream_t stream) {
|
||||
printf("\n BeamHypotheses of size %zu: \n", beam_hyps.size());
|
||||
for (size_t i = 0; i < beam_hyps.size(); i++) {
|
||||
printf("\n [%zu]:\n", i);
|
||||
DumpBeamSearchScorer<<<1, 1, 0, stream>>>(beam_hyps[i]);
|
||||
cudaDeviceSynchronize();
|
||||
}
|
||||
}
|
||||
#endif
|
||||
|
||||
void LaunchBeamSearchScorer_Process(BeamScorerState& state_cpu,
|
||||
BeamScorerState& state,
|
||||
gsl::span<const int32_t> sequences,
|
||||
|
|
@ -431,18 +464,46 @@ void LaunchBeamSearchScorer_Process(BeamScorerState& state_cpu,
|
|||
gsl::span<const int32_t> next_tokens,
|
||||
gsl::span<const int32_t> next_indices,
|
||||
cudaStream_t stream) {
|
||||
BeamSearchScorer_Process<<<1, state_cpu.batch_size_, 0, stream>>>(state_cpu,
|
||||
state,
|
||||
sequences.data(),
|
||||
sequence_length,
|
||||
beam_hyps.data(),
|
||||
next_beam_scores.data(),
|
||||
next_beam_tokens.data(),
|
||||
next_beam_indices.data(),
|
||||
hypothesis_buffer.data(),
|
||||
next_scores.data(),
|
||||
next_tokens.data(),
|
||||
next_indices.data());
|
||||
#ifdef DEBUG_GENERATION
|
||||
printf("\n Before BeamSearchScorer_Process: \n");
|
||||
DumpBeamHypotheses(beam_hyps, stream);
|
||||
DumpBeamScorerStates(state_cpu, state, stream);
|
||||
#endif
|
||||
|
||||
if (state_cpu.early_stopping_) {
|
||||
BeamSearchScorer_Process<true><<<1, state_cpu.batch_size_, 0, stream>>>(state_cpu,
|
||||
state,
|
||||
sequences.data(),
|
||||
sequence_length,
|
||||
beam_hyps.data(),
|
||||
next_beam_scores.data(),
|
||||
next_beam_tokens.data(),
|
||||
next_beam_indices.data(),
|
||||
hypothesis_buffer.data(),
|
||||
next_scores.data(),
|
||||
next_tokens.data(),
|
||||
next_indices.data());
|
||||
} else {
|
||||
BeamSearchScorer_Process<false><<<1, state_cpu.batch_size_, 0, stream>>>(state_cpu,
|
||||
state,
|
||||
sequences.data(),
|
||||
sequence_length,
|
||||
beam_hyps.data(),
|
||||
next_beam_scores.data(),
|
||||
next_beam_tokens.data(),
|
||||
next_beam_indices.data(),
|
||||
hypothesis_buffer.data(),
|
||||
next_scores.data(),
|
||||
next_tokens.data(),
|
||||
next_indices.data());
|
||||
}
|
||||
|
||||
#ifdef DEBUG_GENERATION
|
||||
cudaDeviceSynchronize();
|
||||
printf("\n After BeamSearchScorer_Process: \n");
|
||||
DumpBeamHypotheses(beam_hyps, stream);
|
||||
DumpBeamScorerStates(state_cpu, state, stream);
|
||||
#endif
|
||||
}
|
||||
|
||||
__global__ void BeamSearchScorer_AppendNextTokenToSequences1(BeamScorerState& state,
|
||||
|
|
@ -600,14 +661,14 @@ template <typename T>
|
|||
void LaunchBeamSearchScoreCopy(gsl::span<const float> final_scores,
|
||||
gsl::span<T> output_scores,
|
||||
cudaStream_t stream) {
|
||||
ORT_ENFORCE(final_scores.size() == output_scores.size());
|
||||
constexpr unsigned ThreadPerBlock = 256;
|
||||
unsigned num_blocks = (unsigned)((final_scores.size() + (ThreadPerBlock - 1))/ ThreadPerBlock);
|
||||
ORT_ENFORCE(final_scores.size() == output_scores.size());
|
||||
constexpr unsigned ThreadPerBlock = 256;
|
||||
unsigned num_blocks = (unsigned)((final_scores.size() + (ThreadPerBlock - 1)) / ThreadPerBlock);
|
||||
|
||||
typedef typename ToCudaType<float>::MappedType CudaT;
|
||||
typedef typename ToCudaType<float>::MappedType CudaT;
|
||||
|
||||
FloatConvertAndCopyKernel<<<num_blocks, ThreadPerBlock, 0, stream>>>(
|
||||
final_scores.data(), (CudaT*)output_scores.data(), final_scores.size());
|
||||
FloatConvertAndCopyKernel<<<num_blocks, ThreadPerBlock, 0, stream>>>(
|
||||
final_scores.data(), (CudaT*)output_scores.data(), final_scores.size());
|
||||
}
|
||||
|
||||
template void LaunchBeamSearchScoreCopy(gsl::span<const float> final_scores,
|
||||
|
|
@ -1444,15 +1505,14 @@ void ReorderPastStatesKernelLauncher(void* out_buffer,
|
|||
|
||||
template <typename T>
|
||||
__global__ void CopyCrossQKSingleDecodeStepKernel(
|
||||
T* target, // shape [batchxbeam, layer_head_pair_count, max_length, frame]
|
||||
T* target, // shape [batchxbeam, layer_head_pair_count, max_length, frame]
|
||||
T** qk_layer_pointers,
|
||||
int token_index,
|
||||
int num_layers,
|
||||
int num_heads,
|
||||
const int* cross_qk_layer_head_pairs,
|
||||
int frames,
|
||||
int max_length
|
||||
) {
|
||||
int max_length) {
|
||||
const int pair = blockIdx.x;
|
||||
const int layer_head_pair_count = gridDim.x;
|
||||
const int bbm = blockIdx.y;
|
||||
|
|
@ -1464,7 +1524,7 @@ __global__ void CopyCrossQKSingleDecodeStepKernel(
|
|||
T* src = qk_layer_pointers[layer] + ((int64_t)bbm * num_heads + head) * frames;
|
||||
|
||||
for (int tid = threadIdx.x; tid < frames; tid += blockDim.x) {
|
||||
target[tid] = src[tid]; // use vectorized read write in future if needed
|
||||
target[tid] = src[tid]; // use vectorized read write in future if needed
|
||||
}
|
||||
}
|
||||
|
||||
|
|
@ -1479,8 +1539,7 @@ void LaunchCopyCrossQKSingleDecodeStep(
|
|||
int cross_qk_layer_head_pair_count,
|
||||
const int* cross_qk_layer_head_pairs,
|
||||
int frames,
|
||||
int max_length
|
||||
) {
|
||||
int max_length) {
|
||||
dim3 block(512);
|
||||
dim3 grid(cross_qk_layer_head_pair_count, batchxbeam);
|
||||
typedef typename ToCudaType<float>::MappedType CudaT;
|
||||
|
|
@ -1493,11 +1552,9 @@ void LaunchCopyCrossQKSingleDecodeStep(
|
|||
num_heads,
|
||||
cross_qk_layer_head_pairs,
|
||||
frames,
|
||||
max_length
|
||||
);
|
||||
max_length);
|
||||
}
|
||||
|
||||
|
||||
template <typename T>
|
||||
__global__ void CopyDecoderCrossQKAllStepsKernel(
|
||||
int context_decoding_len,
|
||||
|
|
@ -1505,11 +1562,10 @@ __global__ void CopyDecoderCrossQKAllStepsKernel(
|
|||
int num_return_sequences,
|
||||
int max_length,
|
||||
int frames_of_k,
|
||||
const T* cross_qk_buffer_data, // [batch, num_beams, layer_head_pair_count, max_length, frames]
|
||||
T* cross_qk_output, // [batch, num_return_sequences, layer_head_pair_count, total_decoding_length, frames]
|
||||
const int* cache_indir_data, // [batch, num_beams, max_length]
|
||||
const int32_t* beam_indices
|
||||
) {
|
||||
const T* cross_qk_buffer_data, // [batch, num_beams, layer_head_pair_count, max_length, frames]
|
||||
T* cross_qk_output, // [batch, num_return_sequences, layer_head_pair_count, total_decoding_length, frames]
|
||||
const int* cache_indir_data, // [batch, num_beams, max_length]
|
||||
const int32_t* beam_indices) {
|
||||
const int pair = blockIdx.y;
|
||||
const int layer_head_pair_count = gridDim.y;
|
||||
const int total_decoding_length = gridDim.x;
|
||||
|
|
@ -1522,15 +1578,15 @@ __global__ void CopyDecoderCrossQKAllStepsKernel(
|
|||
const int src_beam = beam_indices[batch * num_beams + ret_seq_id] % num_beams;
|
||||
|
||||
const int64_t offset_in_cache = ((int64_t)batch * num_beams + src_beam) * max_length + token_decoding_index + context_decoding_len;
|
||||
int bm_mapped = ((num_beams <= 1) ? 0: ((token_decoding_index == total_decoding_length - 1) ? ret_seq_id : cache_indir_data[offset_in_cache]));
|
||||
int bm_mapped = ((num_beams <= 1) ? 0 : ((token_decoding_index == total_decoding_length - 1) ? ret_seq_id : cache_indir_data[offset_in_cache]));
|
||||
int bi_src = batch * num_beams + bm_mapped;
|
||||
|
||||
T* target = cross_qk_output +
|
||||
(((int64_t)br * layer_head_pair_count + (int64_t)pair) * total_decoding_length + token_decoding_index) * frames_of_k;
|
||||
T* target = cross_qk_output +
|
||||
(((int64_t)br * layer_head_pair_count + (int64_t)pair) * total_decoding_length + token_decoding_index) * frames_of_k;
|
||||
const T* src = cross_qk_buffer_data +
|
||||
((int64_t)bi_src * layer_head_pair_count * max_length + (int64_t)pair * max_length + token_decoding_index) * frames_of_k;
|
||||
((int64_t)bi_src * layer_head_pair_count * max_length + (int64_t)pair * max_length + token_decoding_index) * frames_of_k;
|
||||
for (int tid = threadIdx.x; tid < frames_of_k; tid += blockDim.x) {
|
||||
target[tid] = src[tid]; // use vectorized read write in future if needed
|
||||
target[tid] = src[tid]; // use vectorized read write in future if needed
|
||||
}
|
||||
}
|
||||
|
||||
|
|
@ -1548,8 +1604,7 @@ void LaunchFinalizeCrossQK(
|
|||
float* cross_qk_output,
|
||||
int num_return_sequences,
|
||||
const int* cache_indir_data,
|
||||
const int32_t* beam_indices
|
||||
) {
|
||||
const int32_t* beam_indices) {
|
||||
int64_t br = (int64_t)batch_size * num_return_sequences;
|
||||
ORT_ENFORCE(br < 65536L && cross_qk_layer_head_pair_count < 65536);
|
||||
const int total_decoding_length = iteration_number - 1;
|
||||
|
|
@ -1558,15 +1613,15 @@ void LaunchFinalizeCrossQK(
|
|||
typedef typename ToCudaType<float>::MappedType CudaT;
|
||||
|
||||
CopyDecoderCrossQKAllStepsKernel<<<grid, block, 0, stream>>>(
|
||||
context_decoding_len,
|
||||
num_beams,
|
||||
num_return_sequences,
|
||||
max_length,
|
||||
frames_of_k,
|
||||
(const CudaT*)cross_qk_buffer_data,
|
||||
(CudaT*)cross_qk_output,
|
||||
cache_indir_data,
|
||||
beam_indices);
|
||||
context_decoding_len,
|
||||
num_beams,
|
||||
num_return_sequences,
|
||||
max_length,
|
||||
frames_of_k,
|
||||
(const CudaT*)cross_qk_buffer_data,
|
||||
(CudaT*)cross_qk_output,
|
||||
cache_indir_data,
|
||||
beam_indices);
|
||||
}
|
||||
|
||||
template <int ElementsPerThreads>
|
||||
|
|
@ -1575,12 +1630,11 @@ __global__ void ForceDecodingIdsKernel(
|
|||
const int vocab_size,
|
||||
const int32_t* force_ids,
|
||||
int id_len,
|
||||
int step
|
||||
) {
|
||||
int step) {
|
||||
const int num_beams = gridDim.y;
|
||||
const int beam = blockIdx.y;
|
||||
const int batch = blockIdx.z;
|
||||
beam_scores += (((int64_t)batch * num_beams + beam)* vocab_size); // move to (batch, beam)
|
||||
beam_scores += (((int64_t)batch * num_beams + beam) * vocab_size); // move to (batch, beam)
|
||||
const int32_t id_wanted = force_ids[((int64_t)batch * id_len) + step];
|
||||
if (id_wanted < 0 || id_wanted >= vocab_size) return;
|
||||
|
||||
|
|
@ -1588,7 +1642,7 @@ __global__ void ForceDecodingIdsKernel(
|
|||
const int32_t block_start_id = blockIdx.x * elements_per_block;
|
||||
|
||||
int32_t token_id = block_start_id + (int)threadIdx.x;
|
||||
#pragma unroll
|
||||
#pragma unroll
|
||||
for (int elem = 0; elem < ElementsPerThreads; elem++) {
|
||||
if (token_id < vocab_size) {
|
||||
beam_scores[token_id] = ((token_id == id_wanted) ? 0.0f : cub::FpLimits<float>::Lowest());
|
||||
|
|
@ -1597,7 +1651,6 @@ __global__ void ForceDecodingIdsKernel(
|
|||
}
|
||||
}
|
||||
|
||||
|
||||
void LaunchForceDecodingIds(
|
||||
float* beam_scores,
|
||||
const int batch_size,
|
||||
|
|
@ -1606,15 +1659,13 @@ void LaunchForceDecodingIds(
|
|||
const int32_t* force_ids,
|
||||
int id_len,
|
||||
int step,
|
||||
cudaStream_t stream
|
||||
) {
|
||||
cudaStream_t stream) {
|
||||
dim3 blocks(512);
|
||||
constexpr int ElementsPerThreads = 4;
|
||||
unsigned gridx = static_cast<unsigned>((vocab_size + 512 * ElementsPerThreads - 1) / (512 * ElementsPerThreads));
|
||||
dim3 grids(gridx, num_beams, batch_size);
|
||||
ForceDecodingIdsKernel<ElementsPerThreads><<<grids, blocks, 0, stream>>>(
|
||||
beam_scores, vocab_size, force_ids, id_len, step
|
||||
);
|
||||
beam_scores, vocab_size, force_ids, id_len, step);
|
||||
}
|
||||
|
||||
template <typename T>
|
||||
|
|
@ -1624,8 +1675,7 @@ __global__ void SaveNoSpeechProbsKernel(
|
|||
const int batch_size,
|
||||
const int num_beams,
|
||||
const int vocab_size,
|
||||
const int no_speech_token_id
|
||||
) {
|
||||
const int no_speech_token_id) {
|
||||
int b = blockIdx.x * blockDim.x + threadIdx.x;
|
||||
if (b < batch_size) {
|
||||
int64_t src_offset = b * num_beams * vocab_size + no_speech_token_id;
|
||||
|
|
@ -1635,20 +1685,19 @@ __global__ void SaveNoSpeechProbsKernel(
|
|||
|
||||
template <typename T>
|
||||
void LaunchSaveNoSpeechProbs(
|
||||
T* result_no_speech_probs, /* [batch]*/
|
||||
const float* probs, /* [batch, num_beams, vocab_size]*/
|
||||
T* result_no_speech_probs, /* [batch]*/
|
||||
const float* probs, /* [batch, num_beams, vocab_size]*/
|
||||
const int batch_size,
|
||||
const int num_beams,
|
||||
const int vocab_size,
|
||||
const int no_speech_token_id,
|
||||
cudaStream_t stream
|
||||
) {
|
||||
cudaStream_t stream) {
|
||||
int tpb = 256;
|
||||
int bpg = (batch_size + 255) / 256;
|
||||
|
||||
typedef typename ToCudaType<T>::MappedType CudaT;
|
||||
SaveNoSpeechProbsKernel<CudaT><<<bpg, tpb, 0, stream>>>(
|
||||
(CudaT*)result_no_speech_probs, probs, batch_size, num_beams, vocab_size, no_speech_token_id);
|
||||
(CudaT*)result_no_speech_probs, probs, batch_size, num_beams, vocab_size, no_speech_token_id);
|
||||
}
|
||||
|
||||
template void LaunchSaveNoSpeechProbs<float>(
|
||||
|
|
@ -1658,8 +1707,7 @@ template void LaunchSaveNoSpeechProbs<float>(
|
|||
const int num_beams,
|
||||
const int vocab_size,
|
||||
const int no_speech_token_id,
|
||||
cudaStream_t stream
|
||||
);
|
||||
cudaStream_t stream);
|
||||
|
||||
template void LaunchSaveNoSpeechProbs<MLFloat16>(
|
||||
MLFloat16* result_no_speech_probs,
|
||||
|
|
@ -1668,8 +1716,7 @@ template void LaunchSaveNoSpeechProbs<MLFloat16>(
|
|||
const int num_beams,
|
||||
const int vocab_size,
|
||||
const int no_speech_token_id,
|
||||
cudaStream_t stream
|
||||
);
|
||||
cudaStream_t stream);
|
||||
|
||||
} // namespace cuda
|
||||
} // namespace contrib
|
||||
|
|
|
|||
|
|
@ -6,6 +6,8 @@
|
|||
#include <stdint.h>
|
||||
#include <cuda_fp16.h>
|
||||
#include <curand_kernel.h>
|
||||
#include <cstdio>
|
||||
#include "contrib_ops/cpu/transformers/generation_shared.h"
|
||||
|
||||
namespace onnxruntime {
|
||||
namespace contrib {
|
||||
|
|
@ -49,6 +51,21 @@ struct HypothesisScore {
|
|||
const int32_t* hypothesis;
|
||||
int hypothesis_length;
|
||||
float score;
|
||||
|
||||
#ifdef DEBUG_GENERATION
|
||||
__device__ void Print() const {
|
||||
printf("HypothesisScore (hypothesis_length=%d, score=%f) \n", hypothesis_length, score);
|
||||
printf(" hypothesis:");
|
||||
if (hypothesis_length > 0 && hypothesis != NULL) {
|
||||
for (int i = 0; i < hypothesis_length; ++i) {
|
||||
printf("%d ", hypothesis[i]);
|
||||
}
|
||||
} else {
|
||||
printf("(empty)");
|
||||
}
|
||||
printf("\n");
|
||||
}
|
||||
#endif
|
||||
};
|
||||
|
||||
struct BeamHypotheses {
|
||||
|
|
@ -71,6 +88,22 @@ struct BeamHypotheses {
|
|||
int pad_token_id, // pad token
|
||||
int32_t* sequences, // buffer with pad token, shape (num_return_sequences, max_length)
|
||||
T* sequences_scores); // buffer for sequence scores, with shape (num_return_sequences)
|
||||
|
||||
#ifdef DEBUG_GENERATION
|
||||
__device__ void Print() const {
|
||||
printf("BeamHypotheses:\n");
|
||||
printf(" beams_count: %d\n", beams_count_);
|
||||
printf(" beams_used: %d\n", beams_used_);
|
||||
printf(" length_penalty: %f\n", length_penalty_);
|
||||
printf(" done: %s\n", done_ ? "true" : "false");
|
||||
|
||||
printf(" beams:\n");
|
||||
for (int i = 0; i < beams_used_; ++i) {
|
||||
printf(" Beam %d:\n", i + 1);
|
||||
beams_[i].Print();
|
||||
}
|
||||
}
|
||||
#endif
|
||||
};
|
||||
|
||||
struct BeamScorerState {
|
||||
|
|
@ -81,9 +114,23 @@ struct BeamScorerState {
|
|||
int pad_token_id_;
|
||||
int eos_token_id_;
|
||||
bool early_stopping_;
|
||||
int not_done_count_; // When zero, every batch entry is done (starts at batch_size_)
|
||||
|
||||
int not_done_count_; // When zero, every batch entry is done (starts at batch_size_)
|
||||
int hypothesis_buffer_used_; // Offset of available buffer, or length of used buffer.
|
||||
|
||||
#ifdef DEBUG_GENERATION
|
||||
__host__ __device__ void Print(bool is_cpu) const {
|
||||
printf("BeamScorerState (cpu=%d) Dump:\n", is_cpu ? 1 : 0);
|
||||
printf(" batch_size_: %d\n", batch_size_);
|
||||
printf(" num_beams_: %d\n", num_beams_);
|
||||
printf(" max_length_: %d\n", max_length_);
|
||||
printf(" num_return_sequences_: %d\n", num_return_sequences_);
|
||||
printf(" pad_token_id_: %d\n", pad_token_id_);
|
||||
printf(" eos_token_id_: %d\n", eos_token_id_);
|
||||
printf(" early_stopping_: %s\n", early_stopping_ ? "true" : "false");
|
||||
printf(" not_done_count_: %d\n", not_done_count_);
|
||||
printf(" hypothesis_buffer_used_: %d\n", hypothesis_buffer_used_);
|
||||
}
|
||||
#endif
|
||||
};
|
||||
|
||||
void LaunchInitializeBeamHypotheses(gsl::span<BeamHypotheses> beam_hyps, float length_penalty, gsl::span<HypothesisScore> beams, int num_beams, cudaStream_t stream);
|
||||
|
|
|
|||
|
|
@ -1090,7 +1090,14 @@ common::Status InferenceSession::Load(const void* model_data, int model_data_len
|
|||
|
||||
const bool strict_shape_type_inference = session_options_.config_options.GetConfigOrDefault(
|
||||
kOrtSessionOptionsConfigStrictShapeTypeInference, "0") == "1";
|
||||
return onnxruntime::Model::Load(std::move(model_proto), PathString(), model,
|
||||
|
||||
std::string external_data_folder_path = session_options_.config_options.GetConfigOrDefault(
|
||||
kOrtSessionOptionsModelExternalInitializersFileFolderPath, "");
|
||||
if (!external_data_folder_path.empty() && model_location_.empty()) {
|
||||
model_location_ = ToPathString(external_data_folder_path + "/virtual_model.onnx");
|
||||
}
|
||||
|
||||
return onnxruntime::Model::Load(std::move(model_proto), model_location_, model,
|
||||
HasLocalSchema() ? &custom_schema_registries_ : nullptr, *session_logger_,
|
||||
ModelOptions(true, strict_shape_type_inference));
|
||||
};
|
||||
|
|
@ -1120,8 +1127,15 @@ common::Status InferenceSession::LoadOnnxModel(ModelProto model_proto) {
|
|||
#endif
|
||||
const bool strict_shape_type_inference = session_options_.config_options.GetConfigOrDefault(
|
||||
kOrtSessionOptionsConfigStrictShapeTypeInference, "0") == "1";
|
||||
|
||||
std::string external_data_folder_path = session_options_.config_options.GetConfigOrDefault(
|
||||
kOrtSessionOptionsModelExternalInitializersFileFolderPath, "");
|
||||
if (!external_data_folder_path.empty() && model_location_.empty()) {
|
||||
model_location_ = ToPathString(external_data_folder_path + "/virtual_model.onnx");
|
||||
}
|
||||
|
||||
// This call will move model_proto to the constructed model instance
|
||||
return onnxruntime::Model::Load(std::move(model_proto), PathString(), model,
|
||||
return onnxruntime::Model::Load(std::move(model_proto), model_location_, model,
|
||||
HasLocalSchema() ? &custom_schema_registries_ : nullptr, *session_logger_,
|
||||
ModelOptions(true, strict_shape_type_inference));
|
||||
};
|
||||
|
|
@ -1157,7 +1171,14 @@ common::Status InferenceSession::Load(std::istream& model_istream, bool allow_re
|
|||
kOrtSessionOptionsConfigStrictShapeTypeInference, "0") == "1";
|
||||
ModelOptions model_opts(allow_released_opsets_only,
|
||||
strict_shape_type_inference);
|
||||
return onnxruntime::Model::Load(std::move(model_proto), PathString(), model,
|
||||
|
||||
std::string external_data_folder_path = session_options_.config_options.GetConfigOrDefault(
|
||||
kOrtSessionOptionsModelExternalInitializersFileFolderPath, "");
|
||||
if (!external_data_folder_path.empty() && model_location_.empty()) {
|
||||
model_location_ = ToPathString(external_data_folder_path + "/virtual_model.onnx");
|
||||
}
|
||||
|
||||
return onnxruntime::Model::Load(std::move(model_proto), model_location_, model,
|
||||
HasLocalSchema() ? &custom_schema_registries_ : nullptr,
|
||||
*session_logger_, model_opts);
|
||||
};
|
||||
|
|
|
|||
|
|
@ -380,7 +380,7 @@ class MinMaxCalibrater(CalibraterBase):
|
|||
else:
|
||||
raise ValueError(
|
||||
f"Unable to guess tensor type for tensor {tensor_name!r}, "
|
||||
f"running shape inference before quantization may resolve this issue."
|
||||
"running shape inference before quantization may resolve this issue."
|
||||
)
|
||||
|
||||
# Include axes in reduce_op when per_channel, always keeping axis=1
|
||||
|
|
@ -1177,6 +1177,7 @@ def create_calibrator(
|
|||
augmented_model_path="augmented_model.onnx",
|
||||
calibrate_method=CalibrationMethod.MinMax,
|
||||
use_external_data_format=False,
|
||||
providers=None,
|
||||
extra_options={}, # noqa: B006
|
||||
):
|
||||
calibrator = None
|
||||
|
|
@ -1243,6 +1244,8 @@ def create_calibrator(
|
|||
|
||||
if calibrator:
|
||||
calibrator.augment_graph()
|
||||
if providers:
|
||||
calibrator.execution_providers = providers
|
||||
calibrator.create_inference_session()
|
||||
return calibrator
|
||||
|
||||
|
|
|
|||
|
|
@ -53,6 +53,7 @@ def get_qnn_qdq_config(
|
|||
weight_symmetric: bool | None = None,
|
||||
keep_removable_activations: bool = False,
|
||||
stride: int | None = None,
|
||||
calibration_providers: list[str] | None = None,
|
||||
) -> StaticQuantConfig:
|
||||
"""
|
||||
Returns a static quantization configuration suitable for running QDQ models on QNN EP.
|
||||
|
|
@ -117,6 +118,8 @@ def get_qnn_qdq_config(
|
|||
are automatically removed if activations are asymmetrically quantized. Keeping these activations
|
||||
is necessary if optimizations or EP transformations will later remove
|
||||
QuantizeLinear/DequantizeLinear operators from the model.
|
||||
calibration_providers: Execution providers to run the session during calibration. Default is None which uses
|
||||
[ "CPUExecutionProvider" ].
|
||||
|
||||
Returns:
|
||||
A StaticQuantConfig object
|
||||
|
|
@ -192,6 +195,7 @@ def get_qnn_qdq_config(
|
|||
op_types_to_quantize=list(op_types.difference(OP_TYPES_TO_EXCLUDE)),
|
||||
per_channel=per_channel,
|
||||
use_external_data_format=(model_has_external_data or model.ByteSize() >= MODEL_SIZE_THRESHOLD),
|
||||
calibration_providers=calibration_providers,
|
||||
extra_options=extra_options,
|
||||
)
|
||||
|
||||
|
|
|
|||
|
|
@ -99,6 +99,7 @@ class StaticQuantConfig(QuantConfig):
|
|||
per_channel=False,
|
||||
reduce_range=False,
|
||||
use_external_data_format=False,
|
||||
calibration_providers=None,
|
||||
extra_options=None,
|
||||
):
|
||||
"""
|
||||
|
|
@ -112,6 +113,8 @@ class StaticQuantConfig(QuantConfig):
|
|||
quant_format: QuantFormat{QOperator, QDQ}.
|
||||
QOperator format quantizes the model with quantized operators directly.
|
||||
QDQ format quantize the model by inserting QuantizeLinear/DeQuantizeLinear on the tensor.
|
||||
calibration_providers: Execution providers to run the session during calibration. Default is None which uses
|
||||
[ "CPUExecutionProvider" ].
|
||||
extra_options:
|
||||
key value pair dictionary for various options in different case. Current used:
|
||||
extra.Sigmoid.nnapi = True/False (Default is False)
|
||||
|
|
@ -219,6 +222,7 @@ class StaticQuantConfig(QuantConfig):
|
|||
self.calibration_data_reader = calibration_data_reader
|
||||
self.calibrate_method = calibrate_method
|
||||
self.quant_format = quant_format
|
||||
self.calibration_providers = calibration_providers
|
||||
self.extra_options = extra_options or {}
|
||||
|
||||
|
||||
|
|
@ -473,6 +477,7 @@ def quantize_static(
|
|||
nodes_to_exclude=None,
|
||||
use_external_data_format=False,
|
||||
calibrate_method=CalibrationMethod.MinMax,
|
||||
calibration_providers=None,
|
||||
extra_options=None,
|
||||
):
|
||||
"""
|
||||
|
|
@ -520,6 +525,8 @@ def quantize_static(
|
|||
List of nodes names to exclude. The nodes in this list will be excluded from quantization
|
||||
when it is not None.
|
||||
use_external_data_format: option used for large size (>2GB) model. Set to False by default.
|
||||
calibration_providers: Execution providers to run the session during calibration. Default is None which uses
|
||||
[ "CPUExecutionProvider" ]
|
||||
extra_options:
|
||||
key value pair dictionary for various options in different case. Current used:
|
||||
extra.Sigmoid.nnapi = True/False (Default is False)
|
||||
|
|
@ -697,6 +704,7 @@ def quantize_static(
|
|||
augmented_model_path=Path(quant_tmp_dir).joinpath("augmented_model.onnx").as_posix(),
|
||||
calibrate_method=calibrate_method,
|
||||
use_external_data_format=use_external_data_format,
|
||||
providers=calibration_providers,
|
||||
extra_options=calib_extra_options,
|
||||
)
|
||||
|
||||
|
|
@ -890,6 +898,7 @@ def quantize(
|
|||
per_channel=quant_config.per_channel,
|
||||
reduce_range=quant_config.reduce_range,
|
||||
use_external_data_format=quant_config.use_external_data_format,
|
||||
calibration_providers=quant_config.calibration_providers,
|
||||
extra_options=quant_config.extra_options,
|
||||
)
|
||||
|
||||
|
|
|
|||
|
|
@ -16,8 +16,8 @@ import subprocess
|
|||
import sys
|
||||
|
||||
TRT_DOCKER_FILES = {
|
||||
"10.7_cuda11.8_cudnn8": "tools/ci_build/github/linux/docker/Dockerfile.ubuntu_cuda11_tensorrt10",
|
||||
"10.7_cuda12.6_cudnn9": "tools/ci_build/github/linux/docker/Dockerfile.ubuntu_cuda12_tensorrt10",
|
||||
"10.8_cuda11.8_cudnn8": "tools/ci_build/github/linux/docker/Dockerfile.ubuntu_cuda11_tensorrt10",
|
||||
"10.8_cuda12.6_cudnn9": "tools/ci_build/github/linux/docker/Dockerfile.ubuntu_cuda12_tensorrt10",
|
||||
"BIN": "tools/ci_build/github/linux/docker/Dockerfile.ubuntu_tensorrt_bin",
|
||||
}
|
||||
|
||||
|
|
|
|||
|
|
@ -423,9 +423,6 @@ TEST(BeamSearchTest, DummyT5WithOuterScopeInitializers) {
|
|||
}
|
||||
|
||||
TEST(BeamSearchTest, DummyT5WithSequenceInputIds) {
|
||||
#if defined(USE_CUDA) && defined(USE_DML)
|
||||
SKIP_CUDA_TEST_WITH_DML;
|
||||
#endif
|
||||
// dummy_t5_with_sequence_input_ids.onnx model generated using following command:
|
||||
// python onnxruntime/test/testdata/dummy_t5_generator.py --output-path dummy_t5_with_sequence_input_ids.onnx --sequence-as-input
|
||||
ModelTester tester(CurrentTestName(), ORT_TSTR("testdata/dummy_t5_with_sequence_input_ids.onnx"));
|
||||
|
|
|
|||
|
|
@ -215,6 +215,52 @@ TEST(CApiTest, TestLoadModelFromArrayWithExternalInitializersFromFileArrayPathRo
|
|||
#endif
|
||||
}
|
||||
|
||||
// The model has external data, Test loading model from array
|
||||
// Extra API required to set the external data path
|
||||
TEST(CApiTest, TestLoadModelFromArrayWithExternalInitializersViaSetExternalDataPath) {
|
||||
std::string model_file_name = "conv_qdq_external_ini.onnx";
|
||||
std::string external_bin_name = "conv_qdq_external_ini.bin";
|
||||
std::string test_folder = "testdata/";
|
||||
std::string model_path = test_folder + model_file_name;
|
||||
std::vector<char> buffer;
|
||||
ReadFileToBuffer(model_path.c_str(), buffer);
|
||||
|
||||
std::vector<char> external_bin_buffer;
|
||||
std::string external_bin_path = test_folder + external_bin_name;
|
||||
ReadFileToBuffer(external_bin_path.c_str(), external_bin_buffer);
|
||||
|
||||
Ort::SessionOptions so;
|
||||
std::string optimized_model_file_name(model_file_name);
|
||||
auto length = optimized_model_file_name.length();
|
||||
optimized_model_file_name.insert(length - 5, "_opt");
|
||||
std::string optimized_file_path(test_folder + optimized_model_file_name);
|
||||
PathString optimized_file_path_t(optimized_file_path.begin(), optimized_file_path.end());
|
||||
|
||||
// Dump the optimized model with external data so that it will unpack the external data from the loaded model
|
||||
so.SetOptimizedModelFilePath(optimized_file_path_t.c_str());
|
||||
|
||||
// set the model external file folder path
|
||||
so.AddConfigEntry(kOrtSessionOptionsModelExternalInitializersFileFolderPath, test_folder.c_str());
|
||||
|
||||
std::string opt_bin_file_name(optimized_model_file_name);
|
||||
opt_bin_file_name.replace(optimized_model_file_name.length() - 4, 4, "bin");
|
||||
so.SetGraphOptimizationLevel(GraphOptimizationLevel::ORT_DISABLE_ALL);
|
||||
so.AddConfigEntry(kOrtSessionOptionsOptimizedModelExternalInitializersFileName, opt_bin_file_name.c_str());
|
||||
so.AddConfigEntry(kOrtSessionOptionsOptimizedModelExternalInitializersMinSizeInBytes, "10");
|
||||
|
||||
Ort::Session session(*ort_env.get(), buffer.data(), buffer.size(), so);
|
||||
|
||||
std::string generated_bin_path = test_folder + opt_bin_file_name;
|
||||
std::vector<char> generated_bin_buffer;
|
||||
ReadFileToBuffer(generated_bin_path.c_str(), generated_bin_buffer);
|
||||
|
||||
ASSERT_EQ(external_bin_buffer, generated_bin_buffer);
|
||||
|
||||
// Cleanup.
|
||||
ASSERT_EQ(std::remove(optimized_file_path.c_str()), 0);
|
||||
ASSERT_EQ(std::remove(generated_bin_path.c_str()), 0);
|
||||
}
|
||||
|
||||
#ifndef _WIN32
|
||||
struct FileDescriptorTraits {
|
||||
using Handle = int;
|
||||
|
|
|
|||
|
|
@ -93,7 +93,9 @@ def _build_aar(args):
|
|||
aar_dir = os.path.join(intermediates_dir, "aar", build_config)
|
||||
jnilibs_dir = os.path.join(intermediates_dir, "jnilibs", build_config)
|
||||
exe_dir = os.path.join(intermediates_dir, "executables", build_config)
|
||||
base_build_command = [sys.executable, BUILD_PY] + build_settings["build_params"] + ["--config=" + build_config]
|
||||
base_build_command = (
|
||||
[sys.executable, BUILD_PY] + build_settings["build_params"] + ["--config=" + build_config, "--use_vcpkg"]
|
||||
)
|
||||
header_files_path = ""
|
||||
|
||||
if qnn_android_build:
|
||||
|
|
|
|||
|
|
@ -71,7 +71,7 @@ jobs:
|
|||
--android_ndk_path $ANDROID_NDK_HOME \
|
||||
--android_abi=x86_64 \
|
||||
--android_api=31 \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--build_shared_lib \
|
||||
--use_qnn static_lib \
|
||||
--qnn_home $(QnnSDKRootDir) \
|
||||
|
|
|
|||
|
|
@ -97,7 +97,7 @@ stages:
|
|||
--android_abi=x86_64 \
|
||||
--android_api=30 \
|
||||
--skip_submodule_sync \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--cmake_generator=Ninja \
|
||||
--build_java
|
||||
displayName: CPU EP, Build and Test
|
||||
|
|
@ -158,7 +158,7 @@ stages:
|
|||
--android_abi=x86_64 \
|
||||
--android_api=29 \
|
||||
--skip_submodule_sync \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--use_nnapi \
|
||||
--build_shared_lib \
|
||||
--cmake_generator=Ninja \
|
||||
|
|
@ -218,7 +218,7 @@ stages:
|
|||
--android_abi=x86_64 \
|
||||
--android_api=29 \
|
||||
--skip_submodule_sync \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--use_nnapi \
|
||||
--build_shared_lib \
|
||||
--cmake_generator=Ninja \
|
||||
|
|
|
|||
|
|
@ -98,7 +98,7 @@ stages:
|
|||
--config Release --update --build \
|
||||
--skip_submodule_sync \
|
||||
--build_shared_lib \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--build_wheel \
|
||||
--enable_onnx_tests --use_cuda --cuda_version=11.8 --cuda_home=/usr/local/cuda-11.8 --cudnn_home=/usr/local/cuda-11.8 \
|
||||
--enable_cuda_profiling \
|
||||
|
|
|
|||
|
|
@ -83,7 +83,7 @@ stages:
|
|||
--config Debug \
|
||||
--skip_submodule_sync \
|
||||
--build_shared_lib \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--enable_onnx_tests --enable_address_sanitizer \
|
||||
--update --build;
|
||||
python3 /onnxruntime_src/tools/ci_build/build.py \
|
||||
|
|
@ -91,7 +91,7 @@ stages:
|
|||
--config Debug \
|
||||
--skip_submodule_sync \
|
||||
--build_shared_lib \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--enable_onnx_tests --enable_address_sanitizer \
|
||||
--test;
|
||||
'
|
||||
|
|
@ -199,7 +199,7 @@ stages:
|
|||
--config Release \
|
||||
--skip_submodule_sync \
|
||||
--build_shared_lib \
|
||||
--parallel --use_binskim_compliant_compile_flags \
|
||||
--parallel --use_vcpkg --use_binskim_compliant_compile_flags \
|
||||
--build_wheel \
|
||||
--build_csharp \
|
||||
--enable_onnx_tests \
|
||||
|
|
|
|||
|
|
@ -67,7 +67,7 @@ jobs:
|
|||
--config Debug Release \
|
||||
--skip_submodule_sync \
|
||||
--build_shared_lib \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--enable_pybind \
|
||||
--enable_onnx_tests \
|
||||
--build_java \
|
||||
|
|
|
|||
|
|
@ -209,7 +209,7 @@ stages:
|
|||
cd /onnxruntime_src/java && /onnxruntime_src/java/gradlew cmakeCheck -DcmakeBuildDir=/build/Release -DUSE_CUDA=1; \
|
||||
cd /tmp; \
|
||||
python3 /onnxruntime_src/tools/ci_build/build.py \
|
||||
--build_dir /build --config Release --test --skip_submodule_sync --build_shared_lib --parallel --use_binskim_compliant_compile_flags --build_wheel --enable_onnx_tests \
|
||||
--build_dir /build --config Release --test --skip_submodule_sync --build_shared_lib --parallel --use_vcpkg --use_binskim_compliant_compile_flags --build_wheel --enable_onnx_tests \
|
||||
--enable_transformers_tool_test --use_cuda --cuda_version=${{parameters.CudaVersion}} --cuda_home=/usr/local/cuda --cudnn_home=/usr/local/cuda \
|
||||
--enable_pybind --build_java --ctest_path "" ; \
|
||||
'
|
||||
|
|
|
|||
|
|
@ -8,10 +8,10 @@ parameters:
|
|||
- name: TrtVersion
|
||||
displayName: TensorRT Version
|
||||
type: string
|
||||
default: 10.7_cuda12.6_cudnn9
|
||||
default: 10.8_cuda12.6_cudnn9
|
||||
values:
|
||||
- 10.7_cuda11.8_cudnn8
|
||||
- 10.7_cuda12.6_cudnn9
|
||||
- 10.8_cuda11.8_cudnn8
|
||||
- 10.8_cuda12.6_cudnn9
|
||||
- BIN
|
||||
|
||||
- name: UseTensorrtOssParser
|
||||
|
|
|
|||
|
|
@ -74,7 +74,7 @@ jobs:
|
|||
--use_qnn $(QnnLibKind) \
|
||||
--qnn_home $(QnnSDKRootDir) \
|
||||
--cmake_generator=Ninja \
|
||||
--update --build --parallel
|
||||
--update --build --parallel --use_vcpkg
|
||||
displayName: Build QNN EP
|
||||
|
||||
- script: |
|
||||
|
|
@ -87,7 +87,7 @@ jobs:
|
|||
--use_qnn $(QnnLibKind) \
|
||||
--qnn_home $(QnnSDKRootDir) \
|
||||
--cmake_generator=Ninja \
|
||||
--test
|
||||
--test --use_vcpkg
|
||||
displayName: Run unit tests
|
||||
|
||||
- task: CmdLine@2
|
||||
|
|
|
|||
|
|
@ -197,7 +197,7 @@ jobs:
|
|||
--enable_nccl \
|
||||
--build_dir /build \
|
||||
--build_shared_lib \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--build_wheel \
|
||||
--skip_submodule_sync \
|
||||
--test --enable_onnx_tests --enable_transformers_tool_test \
|
||||
|
|
|
|||
|
|
@ -58,7 +58,7 @@ jobs:
|
|||
--build_dir build \
|
||||
--skip_submodule_sync \
|
||||
--cmake_generator=Ninja \
|
||||
--parallel --use_binskim_compliant_compile_flags \
|
||||
--parallel --use_vcpkg --use_binskim_compliant_compile_flags \
|
||||
--build_shared_lib \
|
||||
--config Debug \
|
||||
--use_cache \
|
||||
|
|
|
|||
|
|
@ -98,7 +98,7 @@ stages:
|
|||
|
||||
- ${{ else }}:
|
||||
- powershell: |
|
||||
python tools\ci_build\build.py ${{ parameters.BuildCommand }} --use_binskim_compliant_compile_flags --parallel --build_csharp --build --update --config $(BuildConfig) --build_nuget --msbuild_extra_options IncludeMobileTargets=false ${{ variables.build_py_lto_flag }}
|
||||
python tools\ci_build\build.py ${{ parameters.BuildCommand }} --use_binskim_compliant_compile_flags --parallel --use_vcpkg --build_csharp --build --update --config $(BuildConfig) --build_nuget --msbuild_extra_options IncludeMobileTargets=false ${{ variables.build_py_lto_flag }}
|
||||
|
||||
- ${{ if notIn(parameters['sln_platform'], 'Win32', 'x64') }}:
|
||||
# Use cross-compiled protoc
|
||||
|
|
|
|||
|
|
@ -229,7 +229,7 @@ stages:
|
|||
set -e -x
|
||||
export _PYTHON_HOST_PLATFORM=macosx-${{variables.MACOSX_DEPLOYMENT_TARGET}}-universal2
|
||||
python3 -m pip install -r '$(Build.SourcesDirectory)/tools/ci_build/github/linux/docker/scripts/requirements.txt'
|
||||
python3 $(Build.SourcesDirectory)/tools/ci_build/build.py --build_dir $(Build.BinariesDirectory) --use_coreml --skip_submodule_sync --parallel --use_binskim_compliant_compile_flags --config Release --build_wheel ${{ parameters.build_py_parameters }} --use_coreml --cmake_extra_defines CMAKE_OSX_ARCHITECTURES="arm64;x86_64" --update --build
|
||||
python3 $(Build.SourcesDirectory)/tools/ci_build/build.py --build_dir $(Build.BinariesDirectory) --use_coreml --skip_submodule_sync --parallel --use_vcpkg --use_binskim_compliant_compile_flags --config Release --build_wheel ${{ parameters.build_py_parameters }} --use_coreml --cmake_extra_defines CMAKE_OSX_ARCHITECTURES="arm64;x86_64" --update --build
|
||||
displayName: 'Command Line Script'
|
||||
|
||||
- script: |
|
||||
|
|
|
|||
|
|
@ -125,7 +125,7 @@ stages:
|
|||
--cmake_generator "$(VSGenerator)"
|
||||
--enable_pybind
|
||||
--enable_onnx_tests
|
||||
--parallel 8 --use_binskim_compliant_compile_flags --update --build --msvc_toolset 14.40
|
||||
--parallel 8 --use_vcpkg --use_binskim_compliant_compile_flags --update --build --msvc_toolset 14.40
|
||||
$(TelemetryOption) ${{ parameters.BUILD_PY_PARAMETERS }} ${{ parameters.EP_BUILD_FLAGS }} ${{ variables.trt_build_flag }}
|
||||
workingDirectory: '$(Build.BinariesDirectory)'
|
||||
|
||||
|
|
|
|||
|
|
@ -66,7 +66,7 @@ jobs:
|
|||
docker run --rm --volume /data/onnx:/data/onnx:ro --volume $(Build.SourcesDirectory):/onnxruntime_src --volume $(Build.BinariesDirectory):/build \
|
||||
--volume $HOME/.onnx:/home/onnxruntimedev/.onnx -e NIGHTLY_BUILD onnxruntimecpubuildcentos8${{parameters.OnnxruntimeArch}}_packaging /bin/bash -c "python3 \
|
||||
/onnxruntime_src/tools/ci_build/build.py --enable_lto --build_java --build_nodejs --build_dir /build --config Release \
|
||||
--skip_submodule_sync --parallel --use_binskim_compliant_compile_flags --use_vcpkg --build_shared_lib ${{ parameters.AdditionalBuildFlags }} && cd /build/Release && make install DESTDIR=/build/installed"
|
||||
--skip_submodule_sync --parallel --use_vcpkg --use_binskim_compliant_compile_flags --use_vcpkg --build_shared_lib ${{ parameters.AdditionalBuildFlags }} && cd /build/Release && make install DESTDIR=/build/installed"
|
||||
mkdir $(Build.ArtifactStagingDirectory)/testdata
|
||||
cp $(Build.BinariesDirectory)/Release/libcustom_op_library.so* $(Build.ArtifactStagingDirectory)/testdata
|
||||
ls -al $(Build.ArtifactStagingDirectory)
|
||||
|
|
|
|||
|
|
@ -1,7 +1,7 @@
|
|||
variables:
|
||||
common_trt_version: '10.7.0.23'
|
||||
common_trt_version: '10.8.0.43'
|
||||
# As for Debian installation, replace '-1.' by '-1+' when assigning trt version below
|
||||
linux_trt_version_cuda11: ${{ variables.common_trt_version }}-1.cuda11.8
|
||||
linux_trt_version_cuda12: ${{ variables.common_trt_version }}-1.cuda12.6
|
||||
linux_trt_version_cuda12: ${{ variables.common_trt_version }}-1.cuda12.8
|
||||
win_trt_folder_cuda11: TensorRT-${{ variables.common_trt_version }}.Windows10.x86_64.cuda-11.8
|
||||
win_trt_folder_cuda12: TensorRT-${{ variables.common_trt_version }}.Windows10.x86_64.cuda-12.6
|
||||
win_trt_folder_cuda12: TensorRT-${{ variables.common_trt_version }}.Windows10.x86_64.cuda-12.8
|
||||
|
|
@ -13,10 +13,10 @@ parameters:
|
|||
- 12.2
|
||||
- name: TrtVersion
|
||||
type: string
|
||||
default: '10.7.0.23'
|
||||
default: '10.8.0.43'
|
||||
values:
|
||||
- 8.6.1.6
|
||||
- 10.7.0.23
|
||||
- 10.8.0.43
|
||||
|
||||
steps:
|
||||
- ${{ if eq(parameters.DownloadCUDA, true) }}:
|
||||
|
|
@ -42,7 +42,7 @@ steps:
|
|||
- powershell: |
|
||||
Write-Host "##vso[task.setvariable variable=trtCudaVersion;]12.0"
|
||||
displayName: Set trtCudaVersion
|
||||
- ${{ if and(eq(parameters.CudaVersion, '12.2'), eq(parameters.TrtVersion, '10.7.0.23')) }}:
|
||||
- ${{ if and(eq(parameters.CudaVersion, '12.2'), eq(parameters.TrtVersion, '10.8.0.43')) }}:
|
||||
- powershell: |
|
||||
Write-Host "##vso[task.setvariable variable=trtCudaVersion;]12.6"
|
||||
displayName: Set trtCudaVersion
|
||||
|
|
|
|||
|
|
@ -15,10 +15,10 @@ parameters:
|
|||
default: '11.8'
|
||||
- name: win_trt_folder_cuda11
|
||||
type: string
|
||||
default: 'TensorRT-10.7.0.23.Windows10.x86_64.cuda-11.8'
|
||||
default: 'TensorRT-10.8.0.43.Windows10.x86_64.cuda-11.8'
|
||||
- name: win_trt_folder_cuda12
|
||||
type: string
|
||||
default: 'TensorRT-10.7.0.23.Windows10.x86_64.cuda-12.6'
|
||||
default: 'TensorRT-10.8.0.43.Windows10.x86_64.cuda-12.8'
|
||||
|
||||
steps:
|
||||
- ${{ if eq(parameters.DownloadCUDA, 'true') }}:
|
||||
|
|
|
|||
|
|
@ -48,7 +48,7 @@ steps:
|
|||
- script: |
|
||||
set -e -x
|
||||
rm -rf $(Build.BinariesDirectory)/Release
|
||||
python3 $(Build.SourcesDirectory)/tools/ci_build/build.py --update --build ${{ parameters.AdditionalBuildFlags }} --build_dir $(Build.BinariesDirectory) --skip_submodule_sync --parallel --use_binskim_compliant_compile_flags --build_shared_lib --config Release --use_vcpkg
|
||||
python3 $(Build.SourcesDirectory)/tools/ci_build/build.py --update --build ${{ parameters.AdditionalBuildFlags }} --build_dir $(Build.BinariesDirectory) --skip_submodule_sync --parallel --use_vcpkg --use_binskim_compliant_compile_flags --build_shared_lib --config Release --use_vcpkg
|
||||
cd $(Build.BinariesDirectory)/Release
|
||||
make install DESTDIR=$(Build.BinariesDirectory)/installed
|
||||
displayName: 'Build ${{ parameters.MacosArch }}'
|
||||
|
|
|
|||
|
|
@ -98,7 +98,7 @@ jobs:
|
|||
--use_qnn
|
||||
--qnn_home $(QnnSDKRootDir)
|
||||
--enable_pybind
|
||||
--parallel --update
|
||||
--parallel --use_vcpkg --update
|
||||
$(TelemetryOption) ${{ parameters.BUILD_PY_PARAMETERS }}
|
||||
workingDirectory: '$(Build.BinariesDirectory)'
|
||||
|
||||
|
|
|
|||
|
|
@ -83,7 +83,7 @@ jobs:
|
|||
--use_qnn
|
||||
--qnn_home $(QnnSDKRootDir)
|
||||
--enable_pybind
|
||||
--parallel --update --arm64ec
|
||||
--parallel --use_vcpkg --update --arm64ec
|
||||
$(TelemetryOption) ${{ parameters.BUILD_PY_PARAMETERS }}
|
||||
workingDirectory: '$(Build.BinariesDirectory)'
|
||||
|
||||
|
|
|
|||
|
|
@ -76,7 +76,7 @@ jobs:
|
|||
--use_qnn
|
||||
--qnn_home $(QnnSDKRootDir)
|
||||
--enable_pybind
|
||||
--parallel --update
|
||||
--parallel --use_vcpkg --update
|
||||
$(TelemetryOption) ${{ parameters.BUILD_PY_PARAMETERS }}
|
||||
workingDirectory: '$(Build.BinariesDirectory)'
|
||||
|
||||
|
|
|
|||
|
|
@ -89,7 +89,7 @@ jobs:
|
|||
--rocm_home=/opt/rocm \
|
||||
--nccl_home=/opt/rocm \
|
||||
--update \
|
||||
--parallel \
|
||||
--parallel --use_vcpkg \
|
||||
--build_dir /build \
|
||||
--build \
|
||||
--build_wheel \
|
||||
|
|
|
|||
|
|
@ -161,22 +161,10 @@ stages:
|
|||
displayName: 'Generate cmake config'
|
||||
inputs:
|
||||
scriptPath: '$(Build.SourcesDirectory)\tools\ci_build\build.py'
|
||||
arguments: '--config RelWithDebInfo --use_binskim_compliant_compile_flags --enable_lto --disable_rtti --build_dir $(Build.BinariesDirectory) --skip_submodule_sync --build_shared_lib --update --cmake_generator "$(VSGenerator)" --enable_onnx_tests $(TelemetryOption) ${{ parameters.buildparameter }} $(timeoutParameter) $(buildJavaParameter)'
|
||||
arguments: '--config RelWithDebInfo --use_binskim_compliant_compile_flags --enable_lto --disable_rtti --build_dir $(Build.BinariesDirectory) --skip_submodule_sync --build_shared_lib --update --build --use_vcpkg --cmake_generator "$(VSGenerator)" --enable_onnx_tests $(TelemetryOption) ${{ parameters.buildparameter }} $(timeoutParameter) $(buildJavaParameter)'
|
||||
workingDirectory: '$(Build.BinariesDirectory)'
|
||||
|
||||
|
||||
- task: VSBuild@1
|
||||
displayName: 'Build'
|
||||
inputs:
|
||||
solution: '$(Build.BinariesDirectory)\RelWithDebInfo\onnxruntime.sln'
|
||||
platform: ${{ parameters.msbuildPlatform }}
|
||||
configuration: RelWithDebInfo
|
||||
msbuildArchitecture: ${{ parameters.buildArch }}
|
||||
maximumCpuCount: true # default is num logical cores worth of projects building concurrently
|
||||
logProjectEvents: true
|
||||
workingFolder: '$(Build.BinariesDirectory)\RelWithDebInfo'
|
||||
createLogFile: true
|
||||
|
||||
# For CPU job, tests are run in the same machine as building
|
||||
- ${{ if eq(parameters.buildJava, 'true') }}:
|
||||
- template: make_java_win_binaries.yml
|
||||
|
|
|
|||
|
|
@ -106,7 +106,7 @@ stages:
|
|||
parameters:
|
||||
BuildConfig: 'RelWithDebInfo'
|
||||
buildArch: x64
|
||||
additionalBuildFlags: --build_wheel --use_dnnl
|
||||
additionalBuildFlags: --build_wheel --use_dnnl --use_vcpkg
|
||||
msbuildPlatform: x64
|
||||
isX86: false
|
||||
job_name_suffix: x64_release
|
||||
|
|
|
|||
|
|
@ -6,7 +6,7 @@
|
|||
|
||||
# Build base image with required system packages
|
||||
ARG BASEIMAGE=nvidia/cuda:12.5.1-cudnn-devel-ubi8
|
||||
ARG TRT_VERSION=10.7.0.23-1.cuda12.6
|
||||
ARG TRT_VERSION=10.8.0.43-1.cuda12.8
|
||||
FROM $BASEIMAGE AS base
|
||||
ARG TRT_VERSION
|
||||
ENV PATH=/opt/python/cp310-cp310/bin:/usr/local/nvidia/bin:/usr/local/cuda/bin:/usr/src/tensorrt/bin:${PATH}
|
||||
|
|
|
|||
|
|
@ -6,7 +6,7 @@
|
|||
|
||||
# Build base image with required system packages
|
||||
ARG BASEIMAGE=nvidia/cuda:11.8.0-cudnn8-devel-ubi8
|
||||
ARG TRT_VERSION=10.7.0.23-1.cuda11.8
|
||||
ARG TRT_VERSION=10.8.0.43-1.cuda11.8
|
||||
FROM $BASEIMAGE AS base
|
||||
ARG TRT_VERSION
|
||||
ENV PATH=/opt/python/cp310-cp310/bin:/usr/local/nvidia/bin:/usr/local/cuda/bin:/usr/src/tensorrt/bin:${PATH}
|
||||
|
|
|
|||
|
|
@ -6,7 +6,7 @@
|
|||
|
||||
# Build base image with required system packages
|
||||
ARG BASEIMAGE=nvidia/cuda:11.8.0-cudnn8-devel-ubuntu20.04
|
||||
ARG TRT_VERSION=10.7.0.23-1+cuda11.8
|
||||
ARG TRT_VERSION=10.8.0.43-1+cuda11.8
|
||||
ARG LD_LIBRARY_PATH_ARG=/usr/local/lib64:/usr/local/cuda/lib64
|
||||
FROM $BASEIMAGE AS base
|
||||
ARG TRT_VERSION
|
||||
|
|
|
|||
|
|
@ -6,7 +6,7 @@
|
|||
|
||||
# Build base image with required system packages
|
||||
ARG BASEIMAGE=nvidia/cuda:11.8.0-cudnn8-devel-ubuntu22.04
|
||||
ARG TRT_VERSION=10.7.0.23-1+cuda11.8
|
||||
ARG TRT_VERSION=10.8.0.43-1+cuda11.8
|
||||
ARG LD_LIBRARY_PATH_ARG=/usr/local/lib64:/usr/local/cuda/lib64
|
||||
FROM $BASEIMAGE AS base
|
||||
ARG TRT_VERSION
|
||||
|
|
|
|||
|
|
@ -6,7 +6,7 @@
|
|||
|
||||
# Build base image with required system packages
|
||||
ARG BASEIMAGE=nvidia/cuda:11.8.0-cudnn8-devel-ubuntu22.04
|
||||
ARG TRT_VERSION=10.7.0.23-1+cuda11.8
|
||||
ARG TRT_VERSION=10.8.0.43-1+cuda11.8
|
||||
ARG LD_LIBRARY_PATH_ARG=/usr/local/lib64:/usr/local/cuda/lib64
|
||||
FROM $BASEIMAGE AS base
|
||||
ARG TRT_VERSION
|
||||
|
|
|
|||
|
|
@ -31,7 +31,7 @@ RUN pip install --upgrade pip
|
|||
RUN pip install psutil setuptools>=68.2.2
|
||||
|
||||
# Install TensorRT
|
||||
RUN TRT_VERSION="10.7.0.23-1+cuda11.8" &&\
|
||||
RUN TRT_VERSION="10.8.0.43-1+cuda11.8" &&\
|
||||
apt-key adv --fetch-keys https://developer.download.nvidia.com/compute/cuda/repos/ubuntu2004/x86_64/7fa2af80.pub &&\
|
||||
apt-get update &&\
|
||||
apt-get install -y \
|
||||
|
|
|
|||
|
|
@ -31,7 +31,7 @@ RUN pip install --upgrade pip
|
|||
RUN pip install setuptools>=68.2.2 psutil
|
||||
|
||||
# Install TensorRT
|
||||
RUN TRT_VERSION="10.7.0.23-1+cuda12.6" &&\
|
||||
RUN TRT_VERSION="10.8.0.43-1+cuda12.8" &&\
|
||||
apt-key adv --fetch-keys https://developer.download.nvidia.com/compute/cuda/repos/ubuntu2004/x86_64/7fa2af80.pub &&\
|
||||
apt-get update &&\
|
||||
apt-get install -y \
|
||||
|
|
|
|||
|
|
@ -5,7 +5,7 @@
|
|||
ARG BASEIMAGE=nvidia/cuda:11.8.0-cudnn8-devel-ubi8
|
||||
|
||||
FROM $BASEIMAGE
|
||||
ARG TRT_VERSION=10.7.0.23-1.cuda11.8
|
||||
ARG TRT_VERSION=10.8.0.43-1.cuda11.8
|
||||
|
||||
#Install TensorRT only if TRT_VERSION is not empty
|
||||
RUN if [ -n "${TRT_VERSION}" ]; then \
|
||||
|
|
|
|||
|
|
@ -12,7 +12,7 @@ RUN echo "$APT_PREF" > /etc/apt/preferences.d/rocm-pin-600
|
|||
ENV DEBIAN_FRONTEND noninteractive
|
||||
|
||||
RUN apt-get update && \
|
||||
apt-get install -y --no-install-recommends ca-certificates curl libnuma-dev gnupg && \
|
||||
apt-get install -y --no-install-recommends ca-certificates ninja-build git zip curl libnuma-dev gnupg && \
|
||||
curl -sL https://repo.radeon.com/rocm/rocm.gpg.key | apt-key add - &&\
|
||||
printf "deb [arch=amd64] https://repo.radeon.com/rocm/apt/$ROCM_VERSION/ jammy main" | tee /etc/apt/sources.list.d/rocm.list && \
|
||||
printf "deb [arch=amd64] https://repo.radeon.com/amdgpu/$AMDGPU_VERSION/ubuntu jammy main" | tee /etc/apt/sources.list.d/amdgpu.list && \
|
||||
|
|
@ -45,10 +45,10 @@ ENV LANG C.UTF-8
|
|||
WORKDIR /stage
|
||||
|
||||
# Cmake
|
||||
ENV CMAKE_VERSION=3.30.1
|
||||
ENV CMAKE_VERSION=3.31.5
|
||||
RUN cd /usr/local && \
|
||||
wget -q https://github.com/Kitware/CMake/releases/download/v${CMAKE_VERSION}/cmake-${CMAKE_VERSION}-Linux-x86_64.tar.gz && \
|
||||
tar -zxf /usr/local/cmake-3.30.1-Linux-x86_64.tar.gz --strip=1 -C /usr
|
||||
tar -zxf /usr/local/cmake-3.31.5-Linux-x86_64.tar.gz --strip=1 -C /usr
|
||||
|
||||
# ccache
|
||||
RUN mkdir -p /tmp/ccache && \
|
||||
|
|
|
|||
|
|
@ -6,10 +6,10 @@ if exist PATH=%AGENT_TEMPDIRECTORY%\v12.2\ (
|
|||
) else (
|
||||
set PATH=C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.2\bin;C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.2\extras\CUPTI\lib64;%PATH%
|
||||
)
|
||||
set PATH=%AGENT_TEMPDIRECTORY%\TensorRT-10.7.0.23.Windows10.x86_64.cuda-12.6\lib;%PATH%
|
||||
set PATH=%AGENT_TEMPDIRECTORY%\TensorRT-10.8.0.43.Windows10.x86_64.cuda-12.8\lib;%PATH%
|
||||
|
||||
@REM The default version is still cuda v12.2, because set cuda v11.8 after it
|
||||
set PATH=%PATH%;%AGENT_TEMPDIRECTORY%\TensorRT-10.7.0.23.Windows10.x86_64.cuda-11.8\lib
|
||||
set PATH=%PATH%;%AGENT_TEMPDIRECTORY%\TensorRT-10.8.0.43.Windows10.x86_64.cuda-11.8\lib
|
||||
if exist PATH=%AGENT_TEMPDIRECTORY%\v11.8\ (
|
||||
set PATH=%PATH%;%AGENT_TEMPDIRECTORY%\v11.8\bin;%AGENT_TEMPDIRECTORY%\v11.8\extras\CUPTI\lib64
|
||||
) else (
|
||||
|
|
|
|||
|
|
@ -6,6 +6,6 @@ if exist PATH=%AGENT_TEMPDIRECTORY%\v12.2\ (
|
|||
) else (
|
||||
set PATH=%PATH%;C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.2\bin;C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.2\extras\CUPTI\lib64
|
||||
)
|
||||
set PATH=%AGENT_TEMPDIRECTORY%\TensorRT-10.7.0.23.Windows10.x86_64.cuda-12.6\lib;%PATH%
|
||||
set PATH=%AGENT_TEMPDIRECTORY%\TensorRT-10.8.0.43.Windows10.x86_64.cuda-12.8\lib;%PATH%
|
||||
set GRADLE_OPTS=-Dorg.gradle.daemon=false
|
||||
set CUDA_MODULE_LOADING=LAZY
|
||||
|
|
|
|||
Loading…
Reference in a new issue