8000 Update base for Update on "[Inductor] Add a JIT Inductor unit test fo… · pytorch/pytorch@4dde2a9 · GitHub
[go: up one dir, main page]
Skip to content

Commit 4dde2a9

Browse files
desertfiredesertfire
committed
Update base for Update on "[Inductor] Add a JIT Inductor unit test following #146293"
Summary: To follow up #146293, add a JIT Inductor unit test. Other Triton template may need similar fixes. cc voznesenskym penguinwu EikanWang jgong5 Guobing-Chen XiaobingSuper zhuhaozhe blzheng wenzhe-nrv jiayisunx ipiszy yf225 chenyang78 kadeng muchulee8 amjames chauhang aakhundov [ghstack-poisoned]
2 parents 936e0c1 + e01a5e9 commit 4dde2a9

File tree

91 files changed

+3188
-1957
lines changed

Some content is hidden

Large Commits have some content hidden by default. Use the searchbox below for content that may be hidden.

91 files changed

+3188
-1957
lines changed

‎.ci/docker/common/install_onnx.sh

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -31,8 +31,8 @@ pip_install \
3131
pip_install coloredlogs packaging
3232

3333
pip_install onnxruntime==1.18.1
34-
pip_install onnx==1.16.2
35-
pip_install onnxscript==0.1.0.dev20241124 --no-deps
34+
pip_install onnx==1.17.0
35+
pip_install onnxscript==0.1.0 --no-deps
3636
# required by onnxscript
3737
pip_install ml_dtypes
3838

‎.ci/docker/requirements-ci.txt

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -339,7 +339,7 @@ onnx==1.17.0
339339
#Pinned versions:
340340
#test that import:
341341

342-
onnxscript==0.1.0.dev20240817
342+
onnxscript==0.1.0
343343
#Description: Required by mypy and test_public_bindings.py when checking torch.onnx._internal
344344
#Pinned versions:
345345
#test that import:

‎.github/scripts/close_nonexistent_disable_issues.py

Lines changed: 14 additions & 3 deletions
Original file line numberDiff line numberDiff line change
@@ -107,16 +107,20 @@ def close_issue(num: int) -> None:
107107
"Accept": "application/vnd.github.v3+json",
108108
"Authorization": f"token {os.environ['GITHUB_TOKEN']}",
109109
}
110-
requests.post(
110+
response = requests.post(
111111
f"https://api.github.com/repos/pytorch/pytorch/issues/{num}/comments",
112112
data=json.dumps({"body": CLOSING_COMMENT}),
113113
headers=headers,
114114
)
115-
requests.patch(
115+
if response.status_code != 201:
116+
raise RuntimeError(f"Failed to comment on issue {num}: {response.text}")
117+
response = requests.patch(
116118
f"https://api.github.com/repos/pytorch/pytorch/issues/{num}",
117119
data=json.dumps({"state": "closed"}),
118120
headers=headers,
119121
)
122+
if response.status_code != 200:
123+
raise RuntimeError(f"Failed to close issue {num}: {response.text}")
120124

121125

122126
def check_if_exists(
@@ -190,6 +194,13 @@ def check_if_exists(
190194
if args.dry_run:
191195
print("dry run, not actually closing")
192196
else:
197+
failed = False
193198
for item in to_be_closed:
194199
_, (num, _, _) = item
195-
close_issue(num)
200+
try:
201+
close_issue(num)
202+
except RuntimeError as e:
203+
print(e)
204+
failed = True
205+
if failed:
206+
sys.exit(1)

‎.github/workflows/close-nonexistent-disable-issues.yml

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -7,6 +7,8 @@ on:
77
jobs:
88
close-nonexistent-disable-issues:
99
environment: rockset-read-only
10+
permissions:
11+
issues: write
1012
if: github.repository_owner == 'pytorch'
1113
runs-on: ubuntu-latest
1214
steps:

‎aten/src/ATen/Context.h

Lines changed: 0 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -409,11 +409,7 @@ class TORCH_API Context {
409409
bool enabled_cudnnSDP = true;
410410
bool enabled_overrideable = true;
411411
bool allow_fp16_bf16_reduction_mathSDP = false;
412-
#ifdef USE_ROCM
413-
bool benchmark_cudnn = true;
414-
#else
415412
bool benchmark_cudnn = false;
416-
#endif
417413
Float32MatmulPrecision float32_matmul_precision =
418414
c10::utils::check_env("TORCH_ALLOW_TF32_CUBLAS_OVERRIDE") == true
419415
? at::Float32MatmulPrecision::HIGH

‎aten/src/ATen/DeviceAccelerator.cpp

Lines changed: 22 additions & 37 deletions
Original file line numberDiff line numberDiff line change
@@ -5,53 +5,38 @@
55
namespace at::accelerator {
66

77
std::optional<c10::DeviceType> getAccelerator(bool checked) {
8-
// 1. Check PrivateUse1 backends
9-
// We explicitly allow PrivateUse1 and another device at the same time as we
10-
// use this for testing. Whenever a PrivateUse1 device is registered, use it
11-
// first.
12-
// Note that this check is only for hook registration and thus is NOT initializing
13-
// the device or poisoning fork.
14-
if (is_privateuse1_backend_registered()) {
15-
return kPrivateUse1;
8+
#define DETECT_AND_ASSIGN_ACCELERATOR(device_name) \
9+
if (at::has##device_name()) { \
10+
device_type = k##device_name; \
11+
TORCH_CHECK( \
12+
!is_accelerator_detected, \
13+
"Cannot have ", \
14+
device_type.value(), \
15+
" with other accelerators."); \
16+
is_accelerator_detected = true; \
1617
}
1718

18-
// 2. Check runtime backends
19-
// This state is temporary, these runtime checks should be moved to compile-time
20-
// once they provide the new isBuilt API and we are sure they're never in the
21-
// same binary as another accelerator.
22-
#define DETECT_RUNTIME_ACCELERATOR(device_name) \
23-
if (at::has##device_name()) { \
24-
return k##device_name; \
19+
if (is_privateuse1_backend_registered()) {
20+
// We explicitly allow PrivateUse1 and another device at the same time as we
21+
// use this for testing. Whenever a PrivateUse1 device is registered, use it
22+
// first.
23+
return kPrivateUse1;
2524
}
26-
27-
DETECT_RUNTIME_ACCELERATOR(MTIA)
28-
DETECT_RUNTIME_ACCELERATOR(HPU)
29-
30-
#undef DETECT_RUNTIME_ACCELERATOR
31-
32-
// 2. Check compile-time backends
3325
std::optional<c10::DeviceType> device_type = std::nullopt;
34-
35-
#define DETECT_AND_ASSIGN_ACCELERATOR_COMP(device_name) \
36-
if (at::detail::get##device_name##Hooks().isBuilt()) { \
37-
TORCH_CHECK( \
38-
!device_type.has_value(), \
39-
"Cannot have both " #device_name " and ", \
40-
device_type.value(), "."); \
41-
device_type = k##device_name; \
42-
}
43-
44-
DETECT_AND_ASSIGN_ACCELERATOR_COMP(CUDA)
45-
DETECT_AND_ASSIGN_ACCELERATOR_COMP(XPU)
46-
DETECT_AND_ASSIGN_ACCELERATOR_COMP(HIP)
47-
DETECT_AND_ASSIGN_ACCELERATOR_COMP(MPS)
26+
bool is_accelerator_detected = false;
27+
DETECT_AND_ASSIGN_ACCELERATOR(CUDA)
28+
DETECT_AND_ASSIGN_ACCELERATOR(MTIA)
29+
DETECT_AND_ASSIGN_ACCELERATOR(XPU)
30+
DETECT_AND_ASSIGN_ACCELERATOR(HIP)
31+
DETECT_AND_ASSIGN_ACCELERATOR(MPS)
32+
DETECT_AND_ASSIGN_ACCELERATOR(HPU)
4833
if (checked) {
4934
TORCH_CHECK(
5035
device_type, "Cannot access accelerator device when none is available.")
5136
}
5237
return device_type;
5338

54-
#undef DETECT_AND_ASSIGN_ACCELERATOR_COMP
39+
#undef DETECT_AND_ASSIGN_ACCELERATOR
5540
}
5641

5742
bool isAccelerator(c10::DeviceType device_type) {

‎aten/src/ATen/cuda/detail/CUDAHooks.h

Lines changed: 0 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -33,8 +33,6 @@ struct CUDAHooks : public at::CUDAHooksInterface {
3333
bool hasROCM() const override;
3434
const at::cuda::NVRTC& nvrtc() const override;
3535
DeviceIndex current_device() const override;
36-
bool isBuilt() const override {return true;}
37-
bool isAvailable() const override {return hasCUDA();}
3836
bool hasPrimaryContext(DeviceIndex device_index) const override;
3937
Allocator* getCUDADeviceAllocator() const override;
4038
Allocator* getPinnedMemoryAllocator() const override;

‎aten/src/ATen/detail/AcceleratorHooksInterface.h

Lines changed: 0 additions & 17 deletions
Original file line numberDiff line numberDiff line change
@@ -20,23 +20,6 @@ struct TORCH_API AcceleratorHooksInterface {
2020
// squelch -Werror=non-virtual-dtor
2121
virtual ~AcceleratorHooksInterface() = default;
2222

23-
// Whether this backend was enabled at compilation time.
24-
// This function should NEVER throw.
25-
virtual bool isBuilt() const {
26-
return false;
27-
}
28-
29-
// Whether this backend can be used at runtime, meaning it was built,
30-
// its runtime dependencies are available (driver) and at least one
31-
// supported device can be used.
32-
// This function should NEVER throw. This function should NOT initialize the context
33-
// on any device (result of hasPrimaryContext below should not change).
34-
// While it is acceptable for this function to poison fork, it is
35-
// recommended to avoid doing so whenever possible.
36-
virtual bool isAvailable() const {
37-
return false;
38-
}
39-
4023
// Whether the device at device_index is fully initialized or not.
4124
virtual bool hasPrimaryContext(DeviceIndex device_index) const = 0;
4225

‎aten/src/ATen/miopen/Descriptors.h

Lines changed: 4 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -111,10 +111,13 @@ struct ConvolutionDescriptor
111111
&miopenCreateConvolutionDescriptor,
112112
&miopenDestroyConvolutionDescriptor>
113113
{
114-
void set(miopenDataType_t dataType, miopenConvolutionMode_t c_mode, int dim, int* pad, int* stride, int * upscale /* aka dilation */, int groups, bool deterministic) {
114+
void set(miopenDataType_t dataType, miopenConvolutionMode_t c_mode, int dim, int* pad, int* stride, int * upscale /* aka dilation */, int groups, bool benchmark, bool deterministic) {
115115
MIOPEN_CHECK(miopenInitConvolutionNdDescriptor(mut_desc(), dim, pad, stride, upscale, c_mode));
116116
MIOPEN_CHECK(miopenSetConvolutionGroupCount(mut_desc(), groups));
117117
MIOPEN_CHECK(miopenSetConvolutionAttribute(mut_desc(), MIOPEN_CONVOLUTION_ATTRIB_DETERMINISTIC, deterministic ? 1 : 0));
118+
if (benchmark) {
119+
MIOPEN_CHECK(miopenSetConvolutionFindMode(mut_desc(), miopenConvolutionFindModeNormal));
120+
}
118121
}
119122
};
120123

‎aten/src/ATen/miopen/miopen-wrapper.h

Lines changed: 18 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1,3 +1,21 @@
11
#pragma once
22

33
#include <miopen/miopen.h>
4+
#include <miopen/version.h>
5+
6+
#if MIOPEN_VERSION_MAJOR > 3 || (MIOPEN_VERSION_MAJOR == 3 && MIOPEN_VERSION_MINOR >= 4)
7+
// miopen 3.4 moved find mode from private header to public header
8+
#else
9+
// from miopen_internal.h
10+
extern "C" {
11+
12+
typedef enum
13+
{
14+
miopenConvolutionFindModeNormal = 1, /*!< Normal mode */
15+
} miopenConvolutionFindMode_t;
16+
17+
miopenStatus_t miopenSetConvolutionFindMode(
18+
miopenConvolutionDescriptor_t convDesc,
19+
miopenConvolutionFindMode_t findMode);
20+
}
21+
#endif

0 commit comments

Comments
 (0)
0