🔗 Helpful Links

🧪 See artifacts and rendered test results at hud.pytorch.org/pr/144572

• 📄 Preview Python docs built from this PR
• 📄 Preview C++ docs built from this PR
• ❓ Need help or want to give feedback on the CI? Visit the bot commands wiki or our office hours

Note: Links to docs will display an error until the docs builds have been completed.

✅ You can merge normally! (1 Unrelated Failure)

As of commit 0741dce with merge base 863ac20 ():

This comment was automatically generated by Dr. CI and updates every 15 minutes.

Initial questions:
I need opinions on some key issues:

1. To avoid duplicating a lot of code I currently prepare DropoutDescriptor and the memory for it inside the RNNDescriptors constructor. Other solution would be to move memory management into DropoutDescriptor itself, but it would complicate it beyond just being a simple wrapper and we may lose some control over its' internals from outside. Also all the memory management can be moved inside miopen_rnn functions, but that will produce 3 identical code snippets in all of the functions.
2. Struct for device memory pointer. I currently use a simple way of doing it through hipmalloc, but we do almost identical thing in Conv_miopen with Workspace struct. Should they be combined and moved to some utils file?
3. hipmalloc vs c10::hip::HIPCachingAllocator::raw_alloc. I just don't know the difference, what should we use in general situation?
4. RNG seed is currently unused. There is no way to set it from pytorch api. Should we bother with that and give such control to users or stick to setting the seed randomly?

Addressed issues after getting some feedback:

1. Decided to leave DropoutDescriptor usage as is, MIOpen API doesn't really allow for a more elegant solution.
2. Moved memory management into a separate header file for both Conv and RNN files.
3. Direct usage of hipMalloc breaks hipGraph support. Now HIPCachingAllocator is used everywhere.
4. Will be addressed in a separate issue/pr.

@pytorchbot label "topic: not user facing"

@pytorchbot rebase

@pytorchbot started a rebase job onto refs/remotes/origin/viable/strict. Check the current status here

Successfully rebased iupaikov_rnn_dropout_state_fix_upstream onto refs/remotes/origin/viable/strict, please pull locally before adding more changes (for example, via git checkout iupaikov_rnn_dropout_state_fix_upstream && git pull --rebase)

@pytorchbot rebase

@pytorchbot started a rebase job onto refs/remotes/origin/viable/strict. Check the current status here

Successfully rebased iupaikov_rnn_dropout_state_fix_upstream onto refs/remotes/origin/viable/strict, please pull locally before adding more changes (for example, via git checkout iupaikov_rnn_dropout_state_fix_upstream && git pull --rebase)

Thanks everyone extensive code review! Due to the issue with RNG seed usage we decided to postpone this PR to later release. For now I will keep this PR in draft.

The main problem is that when we come down to cpp land for MIOpen API usage we create all of the supporting structs there for each call of the RNN function. This causes RNG machine to generate the same random numbers if we use the same seed. We either will need to keep track of the python context in cpp part of code or have a gpu mem allocated in python and always pass it down.

@pytorchbot rebase

@pytorchbot started a rebase job onto refs/remotes/origin/viable/strict. Check the current status here

Rebase failed due to Command git -C /home/runner/work/pytorch/pytorch rebase refs/remotes/origin/viable/strict pull/144572/head returned non-zero exit code 1

Rebasing (1/7)
Rebasing (2/7)
Auto-merging aten/src/ATen/CMakeLists.txt
Auto-merging aten/src/ATen/native/miopen/Conv_miopen.cpp
CONFLICT (content): Merge conflict in aten/src/ATen/native/miopen/Conv_miopen.cpp
error: could not apply dc9591a80b5... Moved gpu memory management to a separate header utility
hint: Resolve all conflicts manually, mark them as resolved with
hint: "git add/rm <conflicted_files>", then run "git rebase --continue".
hint: You can instead skip this commit: run "git rebase --skip".
hint: To abort and get back to the state before "git rebase", run "git rebase --abort".
hint: Disable this message with "git config set advice.mergeConflict false"
Could not apply dc9591a80b5... Moved gpu memory management to a separate header utility

Raised by https://github.com/pytorch/pytorch/actions/runs/14111718518

I implemented a proper way of dropout state restoration that doesn't require setting a seed to random number to produce correct results for a linked test. For now the seed is always set to 0, but this will later be addressed in a different PR. There are some complications with HIP API regarding random number generator tied to a device.

We decided to store each dropout state in thread_local storage to avoid multithreading issues here. Dropout descriptor needs to also be stored in such way for API to function properly. From initial tests this shouldn't cause any issues or impact performance in major way, but we will continue to monitor this and rework this if needed.

Also since we need to store the dropout state size and later probably some other stuff as well if we need a descriptor for each device I decided to leave a struct approach to manage the workspace memory. Tried the proposed variant but readability of the code suffered in a major way and I changed it back to struct with a more specific naming.

@jeffdaily can we merge this? The issue is just credentials.

@pytorchbot merge -f "flaky test failure on rocm, safe to merge"

Merge started

Your change will be merged immediately since you used the force (-f) flag, bypassing any CI checks (ETA: 1-5 minutes). Please use -f as last resort and instead consider -i/--ignore-current to continue the merge ignoring current failures. This will allow currently pending tests to finish and report signal before the merge.

Learn more about merging in the wiki.

Questions? Feedback? Please reach out to the PyTorch DevX Team