Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Standardize workspace abstraction #2524

Merged
merged 25 commits into from
Dec 16, 2023
Merged
Changes from 1 commit
Commits
Show all changes
25 commits
Select commit Hold shift + click to select a range
5d9d7fc
added checks on workspace params
amberhassaan Nov 1, 2023
75e18bb
addressed review comments
amberhassaan Nov 8, 2023
9859d97
Merge remote-tracking branch 'origin/develop' into amber/add-workspac…
amberhassaan Nov 9, 2023
9bb0750
fix release build warning
amberhassaan Nov 9, 2023
675b6a9
WIP: workspace abstraction
amberhassaan Nov 11, 2023
4b33079
added an abstraction for workspace
amberhassaan Nov 15, 2023
6b12bcf
Merge branch 'develop' into amber/add-workspace-check
amberhassaan Nov 15, 2023
cdb9325
fix a check
amberhassaan Nov 15, 2023
71f709b
fix format
amberhassaan Nov 15, 2023
ae5d50e
missed some instances
amberhassaan Nov 15, 2023
02bf9d0
Merge remote-tracking branch 'origin/develop' into amber/workspace-ab…
amberhassaan Nov 15, 2023
83d622b
bring back the workspace round up logic
amberhassaan Nov 15, 2023
4bd4253
formatting
amberhassaan Nov 15, 2023
6eef5fe
fix hip tidy issues. add more checks
amberhassaan Nov 16, 2023
f22aa2d
fix a bug with zeroing out rnn workspace
amberhassaan Nov 16, 2023
d6ae39f
fix hip-tidy error
amberhassaan Nov 16, 2023
385d816
Merge branch 'amber/add-workspace-check' into amber/workspace-abstrac…
amberhassaan Nov 21, 2023
815b73a
Merge remote-tracking branch 'origin/develop' into amber/workspace-ab…
amberhassaan Nov 21, 2023
98ec151
address review comments
amberhassaan Nov 21, 2023
437ea5c
Merge remote-tracking branch 'origin/develop' into amber/workspace-ab…
amberhassaan Dec 6, 2023
43bb198
Merge branch 'develop' into amber/workspace-abstraction
amberhassaan Dec 7, 2023
0b9a725
Merge branch 'develop' into amber/workspace-abstraction
amberhassaan Dec 12, 2023
9880704
Merge branch 'develop' into amber/workspace-abstraction
amberhassaan Dec 12, 2023
75b9f45
Merge branch 'develop' into amber/workspace-abstraction
amberhassaan Dec 14, 2023
f1e32d7
Merge branch 'develop' into amber/workspace-abstraction
junliume Dec 14, 2023
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
Prev Previous commit
Next Next commit
added an abstraction for workspace
  • Loading branch information
amberhassaan committed Nov 15, 2023

Verified

This commit was created on GitHub.com and signed with GitHub’s verified signature. The key has expired.
commit 4b3307912512433d66417cc28b6b91c13bf22128
131 changes: 50 additions & 81 deletions test/conv_common.hpp

Large diffs are not rendered by default.

2 changes: 0 additions & 2 deletions test/find_2_conv.cpp
Original file line number Diff line number Diff line change
@@ -312,8 +312,6 @@ struct Find2Test : test_driver
{
std::cerr << "Running a solution..." << std::endl;

auto& handle_deref = get_handle();

std::size_t workspace_size;
EXPECT_EQUAL(miopenGetSolutionWorkspaceSize(solution, &workspace_size),
miopenStatusSuccess);
1 change: 1 addition & 0 deletions test/find_db.cpp
Original file line number Diff line number Diff line change
@@ -27,6 +27,7 @@
#include "test.hpp"
#include "driver.hpp"
#include "get_handle.hpp"
#include "workspace.hpp"

#include <miopen/convolution.hpp>
#include <miopen/conv/problem_description.hpp>
68 changes: 37 additions & 31 deletions test/gru_common.hpp
Original file line number Diff line number Diff line change
@@ -35,6 +35,7 @@
#include "verify.hpp"
#include "rnn_util.hpp"
#include "random.hpp"
#include "workspace.hpp"
#include <array>
#include <cmath>
#include <ctime>
@@ -1964,7 +1965,7 @@ struct verify_forward_infer_gru
auto&& handle = get_handle();

size_t out_sz = 0;
size_t workSpaceSize = 0;
size_t workspace_size = 0;

std::vector<miopen::TensorDescriptor> inputCPPDescs;
std::vector<miopenTensorDescriptor_t> inputDescs;
@@ -1979,9 +1980,9 @@ struct verify_forward_infer_gru
hiddenSize * ((dirMode != 0) ? 2 : 1),
miopen::deref(rnnDesc).dataType);

miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workSpaceSize);
miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workspace_size);

std::vector<T> workSpace(workSpaceSize / sizeof(T));
std::vector<T> workSpace(workspace_size / sizeof(T));

auto input_dev = handle.Write(input);

@@ -2028,7 +2029,7 @@ struct verify_forward_infer_gru
&hiddenDesc,
nullptr,
workSpace_dev.get(),
workSpaceSize);
workspace_size);

#if(MIO_GRU_TEST_DEBUG == 2)
auto outdata = handle.Read<T>(output_dev, output.size());
@@ -2249,7 +2250,7 @@ struct verify_forward_train_gru
auto&& handle = get_handle();

size_t out_sz = 0;
size_t workSpaceSize = 0;
size_t workspace_size = 0;
size_t reserveSpaceSize = 0;

std::vector<miopen::TensorDescriptor> inputCPPDescs;
@@ -2265,11 +2266,11 @@ struct verify_forward_train_gru
hiddenSize * ((dirMode != 0) ? 2 : 1),
miopen::deref(rnnDesc).dataType);

miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workSpaceSize);
miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workspace_size);
miopenGetRNNTrainingReserveSize(
&handle, rnnDesc, seqLength, inputDescs.data(), &reserveSpaceSize);

std::vector<T> workSpace(workSpaceSize / sizeof(T));
std::vector<T> workSpace(workspace_size / sizeof(T));
std::vector<T> reserveSpace((reserveSpaceSize + sizeof(T) - 1) / sizeof(T));

auto input_dev = handle.Write(input);
@@ -2319,7 +2320,7 @@ struct verify_forward_train_gru
&hiddenDesc,
nullptr,
workSpace_dev.get(),
workSpaceSize,
workspace_size,
reserveSpace_dev.get(),
reserveSpaceSize);

@@ -2484,7 +2485,7 @@ struct verify_backward_data_gru
int bi = dirMode != 0 ? 2 : 1;
int hy_h = hiddenSize;
int bi_stride = bi * hy_h;
size_t workSpaceSize;
size_t workspace_size;

std::vector<miopen::TensorDescriptor> inputCPPDescs;
std::vector<miopenTensorDescriptor_t> inputDescs;
@@ -2494,8 +2495,8 @@ struct verify_backward_data_gru
// Outputs ----------
size_t in_sz = 0;
miopenGetRNNInputTensorSize(&handle, rnnDesc, seqLength, inputDescs.data(), &in_sz);
miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workSpaceSize);
std::vector<T> workSpace(workSpaceSize / sizeof(T));
miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workspace_size);
std::vector<T> workSpace(workspace_size / sizeof(T));
std::vector<T> dx(in_sz / sizeof(T));
std::vector<T> dhx(initHidden.size());

@@ -2563,7 +2564,6 @@ struct verify_backward_data_gru
auto&& handle = get_handle();

size_t out_sz = 0;
size_t workSpaceSize = 0;

std::vector<miopen::TensorDescriptor> inputCPPDescs;
std::vector<miopenTensorDescriptor_t> inputDescs;
@@ -2578,16 +2578,19 @@ struct verify_backward_data_gru
hiddenSize * ((dirMode != 0) ? 2 : 1),
miopen::deref(rnnDesc).dataType);

miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workSpaceSize);
std::vector<T> workSpace(workSpaceSize / sizeof(T));
auto workSpace_dev = handle.Write(workSpace);
size_t workspace_size = 0;
miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workspace_size);
Workspace wspace{};
wspace.resize(workspace_size);

miopenGetRNNInputTensorSize(&handle, rnnDesc, seqLength, outputDescs.data(), &out_sz);
auto yin_dev = handle.Write(yin);
auto dyin_dev = handle.Write(dy);
auto reserveSpace_dev = handle.Write(reserveSpace);
auto weights_dev = handle.Write(weights);

Workspace rspace{};
rspace.Write(reserveSpace);

std::vector<int> hlens(3, 0);
hlens[0] = nLayers * (dirMode != 0 ? 2 : 1);
hlens[1] = batch_seq[0];
@@ -2633,15 +2636,15 @@ struct verify_backward_data_gru
((nodhx) ? nullptr : dhx_dev.get()),
&hiddenDesc,
nullptr,
workSpace_dev.get(),
workSpaceSize,
reserveSpace_dev.get(),
reserveSpace.size() * sizeof(T));
wspace.ptr(),
wspace.size(),
rspace.ptr(),
rspace.size());

auto retSet = std::make_tuple(handle.Read<T>(dx_dev, dx.size()),
(nodhx ? initHidden : handle.Read<T>(dhx_dev, dhx.size())),
handle.Read<T>(reserveSpace_dev, reserveSpace.size()),
handle.Read<T>(workSpace_dev, workSpace.size()));
rspace.Read<std::vector<T>>(),
wspace.Read<std::vector<T>>());

#if(MIO_RNN_TIME_EVERYTHING == 1)
auto t_end = std::chrono::high_resolution_clock::now();
@@ -2840,8 +2843,11 @@ struct verify_backward_weights_gru
hiddenSize * ((dirMode != 0) ? 2 : 1),
miopen::deref(rnnDesc).dataType);

auto workSpace_dev = handle.Write(workSpace);
auto reserveSpace_dev = handle.Write(reserveSpace);
Workspace wspace{};
wspace.Write(workSpace);
Workspace rspace{};
rspace.Write(reserveSpace);

std::vector<T> dweights(weightSize);
auto dweights_dev = handle.Write(dweights);
miopen::TensorDescriptor weightDesc(miopen::deref(rnnDesc).dataType, {weightSize});
@@ -2869,10 +2875,10 @@ struct verify_backward_weights_gru
dy_dev.get(),
&weightDesc,
dweights_dev.get(),
workSpace_dev.get(),
workSpace.size() * sizeof(T),
reserveSpace_dev.get(),
reserveSpace.size() * sizeof(T));
wspace.ptr(),
wspace.size(),
rspace.ptr(),
rspace.size());

#if(MIO_RNN_TIME_EVERYTHING == 1)
auto t_end = std::chrono::high_resolution_clock::now();
@@ -3123,10 +3129,10 @@ struct gru_basic_driver : test_driver
size_t reserveSpaceSize;
miopenGetRNNTrainingReserveSize(
&handle, rnnDesc, seqLength, inputDescs.data(), &reserveSpaceSize);
size_t workSpaceSize;
miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workSpaceSize);
size_t workspace_size;
miopenGetRNNWorkspaceSize(&handle, rnnDesc, seqLength, inputDescs.data(), &workspace_size);

size_t total_mem = statesSizeInBytes + reserveSpaceSize + workSpaceSize + 2 * out_sz +
size_t total_mem = statesSizeInBytes + reserveSpaceSize + workspace_size + 2 * out_sz +
(in_sz + wei_sz + (nohx ? 0 : hx_sz) + (nohy ? 0 : hx_sz) +
(nodhx ? 0 : hx_sz) + (nodhy ? 0 : hx_sz)) *
sizeof(T);
Loading