From d9585236489c45452a4205f2b00a847876a3cf84 Mon Sep 17 00:00:00 2001 From: Krzysztof Lecki Date: Tue, 26 Oct 2021 17:33:12 +0200 Subject: [PATCH] Rename InputRef/OutputRef to Input/Output in workspace API (#3451) Signed-off-by: Krzysztof Lecki --- dali/benchmark/caffe2_alexnet_bench.cc | 4 +- dali/benchmark/caffe_alexnet_bench.cc | 4 +- dali/benchmark/decoder_bench.cc | 2 +- dali/benchmark/file_reader_alexnet_bench.cc | 2 +- dali/benchmark/resnet50_bench.cc | 6 +- dali/benchmark/resnet50_nvjpeg_bench.cc | 2 +- dali/c_api/c_api.cc | 24 ++++---- dali/c_api/c_api_test.cc | 4 +- .../audio/mel_scale/mel_filter_bank.cc | 6 +- .../audio/mel_scale/mel_filter_bank_gpu.cc | 6 +- dali/operators/audio/mfcc/mfcc.cc | 8 +-- dali/operators/audio/mfcc/mfcc.cu | 6 +- dali/operators/audio/mfcc/mfcc.h | 2 +- dali/operators/audio/nonsilence_op.cc | 2 +- dali/operators/audio/nonsilence_op.h | 8 +-- dali/operators/audio/preemphasis_filter_op.cc | 6 +- dali/operators/audio/preemphasis_filter_op.cu | 6 +- dali/operators/audio/preemphasis_filter_op.h | 2 +- dali/operators/bbox/bb_flip.cc | 6 +- dali/operators/bbox/bb_flip.cu | 4 +- dali/operators/bbox/bb_flip.h | 2 +- dali/operators/bbox/bbox_paste.cc | 4 +- dali/operators/debug/dump_image.cc | 4 +- dali/operators/debug/dump_image.cu | 4 +- .../decoder/audio/audio_decoder_op.cc | 6 +- dali/operators/decoder/host/host_decoder.cc | 4 +- .../nvjpeg/nvjpeg_decoder_decoupled_api.h | 22 +++---- dali/operators/generic/cast.cc | 4 +- dali/operators/generic/cast.cu | 6 +- dali/operators/generic/cast.h | 2 +- dali/operators/generic/constant.cc | 4 +- dali/operators/generic/constant.cu | 2 +- dali/operators/generic/erase/erase.cc | 8 +-- dali/operators/generic/erase/erase.cu | 8 +-- dali/operators/generic/expand_dims.cc | 2 +- dali/operators/generic/flip.cc | 4 +- dali/operators/generic/flip.cu | 4 +- dali/operators/generic/flip.h | 4 +- dali/operators/generic/join.cc | 16 ++--- dali/operators/generic/lookup_table.cc | 4 +- dali/operators/generic/lookup_table.cu | 4 +- dali/operators/generic/lookup_table.h | 2 +- dali/operators/generic/one_hot.cc | 4 +- dali/operators/generic/one_hot.cu | 10 ++-- dali/operators/generic/one_hot.h | 4 +- dali/operators/generic/pad.cc | 6 +- dali/operators/generic/pad.cu | 6 +- dali/operators/generic/pad.h | 4 +- dali/operators/generic/permute_batch.cc | 8 +-- dali/operators/generic/permute_batch.h | 4 +- dali/operators/generic/reduce/mean.h | 2 +- dali/operators/generic/reduce/mean_square.h | 2 +- dali/operators/generic/reduce/reduce.h | 12 ++-- .../generic/reduce/reduce_with_mean_input.h | 16 ++--- .../generic/reduce/root_mean_square.h | 2 +- dali/operators/generic/reduce/sum.h | 2 +- dali/operators/generic/reshape.cc | 16 ++--- dali/operators/generic/reshape.h | 2 +- dali/operators/generic/roi_random_crop.cc | 4 +- dali/operators/generic/shapes.h | 10 ++-- dali/operators/generic/slice/slice_attr.h | 4 +- dali/operators/generic/slice/slice_base.cc | 8 +-- dali/operators/generic/slice/slice_base.cu | 8 +-- dali/operators/generic/slice/slice_base.h | 4 +- dali/operators/generic/slice/subscript.cc | 8 +-- dali/operators/generic/slice/subscript.cu | 4 +- dali/operators/generic/slice/subscript.h | 6 +- dali/operators/generic/squeeze.cc | 2 +- dali/operators/generic/transpose/transpose.cc | 4 +- dali/operators/generic/transpose/transpose.h | 4 +- .../generic/transpose/transpose_gpu.cc | 6 +- .../affine_transforms/combine_transforms.cc | 8 +-- .../affine_transforms/transform_base_op.h | 8 +-- dali/operators/geometry/coord_flip.cc | 6 +- dali/operators/geometry/coord_flip.cu | 4 +- dali/operators/geometry/coord_flip.h | 2 +- dali/operators/geometry/coord_transform.cc | 6 +- dali/operators/geometry/coord_transform.cu | 6 +- dali/operators/geometry/coord_transform.h | 6 +- .../image/color/brightness_contrast.cc | 8 +-- .../image/color/brightness_contrast.cu | 8 +-- .../image/color/brightness_contrast.h | 2 +- .../image/color/color_space_conversion.cc | 6 +- .../image/color/color_space_conversion.cu | 6 +- .../image/color/color_space_conversion.h | 4 +- dali/operators/image/color/color_twist.cc | 6 +- dali/operators/image/color/color_twist.cu | 6 +- dali/operators/image/color/color_twist.h | 2 +- dali/operators/image/color/old_color_twist.cc | 10 ++-- .../image/convolution/gaussian_blur.cc | 8 +-- .../image/convolution/gaussian_blur.cu | 2 +- .../image/convolution/gaussian_blur_gpu.h | 6 +- dali/operators/image/crop/bbox_crop.cc | 16 ++--- dali/operators/image/crop/crop.h | 4 +- .../image/crop/crop_mirror_normalize.cc | 6 +- .../image/crop/crop_mirror_normalize.cu | 6 +- .../image/crop/crop_mirror_normalize.h | 2 +- .../jpeg_compression_distortion_op.h | 4 +- .../jpeg_compression_distortion_op_cpu.cc | 6 +- .../jpeg_compression_distortion_op_gpu.cu | 6 +- dali/operators/image/mask/grid_mask.cc | 8 +-- dali/operators/image/mask/grid_mask.cu | 8 +-- dali/operators/image/paste/multipaste.cc | 4 +- dali/operators/image/paste/multipaste.cu | 6 +- dali/operators/image/paste/multipaste.h | 8 +-- dali/operators/image/paste/paste.cu | 6 +- .../image/peek_shape/peek_image_shape.h | 6 +- .../remap/displacement_filter_impl_cpu.h | 10 ++-- .../remap/displacement_filter_impl_gpu.cuh | 10 ++-- dali/operators/image/remap/rotate_params.h | 2 +- dali/operators/image/remap/warp.h | 16 ++--- .../image/remap/warp_affine_params.h | 4 +- .../image/remap/warp_param_provider.h | 4 +- .../image/resize/random_resized_crop.cc | 6 +- .../image/resize/random_resized_crop.cu | 4 +- .../image/resize/random_resized_crop.h | 2 +- dali/operators/image/resize/resize.cc | 12 ++-- dali/operators/image/resize/resize.h | 2 +- .../image/resize/resize_crop_mirror.h | 8 +-- dali/operators/math/expressions/arithmetic.cc | 4 +- dali/operators/math/expressions/arithmetic.cu | 4 +- dali/operators/math/expressions/arithmetic.h | 6 +- .../math/expressions/arithmetic_test.cc | 30 +++++----- .../expressions/expression_impl_factory.h | 10 ++-- dali/operators/math/normalize/normalize.cc | 6 +- dali/operators/math/normalize/normalize.h | 4 +- .../operators/math/normalize/normalize_gpu.cu | 6 +- dali/operators/numba_function/numba_func.cc | 12 ++-- .../python_function/dltensor_function.cc | 10 ++-- .../python_function/dltensor_function.h | 4 +- dali/operators/random/batch_permutation.cc | 4 +- dali/operators/random/rng_base.h | 10 ++-- dali/operators/random/rng_base_cpu.h | 6 +- dali/operators/random/rng_base_gpu.cuh | 4 +- dali/operators/reader/coco_reader_op.cc | 14 ++--- dali/operators/reader/coco_reader_op_test.cc | 12 ++-- dali/operators/reader/file_reader_op.h | 4 +- dali/operators/reader/nemo_asr_reader_op.cc | 8 +-- .../reader/numpy_reader_gpu_op_impl.cu | 4 +- dali/operators/reader/numpy_reader_op.cc | 2 +- dali/operators/reader/parser/caffe2_parser.h | 14 ++--- dali/operators/reader/parser/caffe_parser.h | 4 +- dali/operators/reader/parser/parser_test.cc | 2 +- .../operators/reader/parser/recordio_parser.h | 4 +- .../reader/parser/sequence_parser.cc | 2 +- .../operators/reader/parser/tfrecord_parser.h | 2 +- dali/operators/reader/reader_op.h | 6 +- dali/operators/reader/reader_op_test.cc | 6 +- .../reader/video_reader_decoder_op.cc | 4 +- .../reader/video_reader_decoder_op_test.cc | 18 +++--- dali/operators/reader/video_reader_op.h | 8 +-- dali/operators/reader/video_reader_op_test.cc | 60 +++++++++---------- dali/operators/reader/webdataset_reader_op.cc | 24 ++++---- .../segmentation/random_mask_pixel.cc | 8 +-- .../segmentation/random_object_bbox.cc | 10 ++-- dali/operators/segmentation/select_masks.cc | 16 ++--- dali/operators/sequence/element_extract.h | 6 +- .../sequence/optical_flow/optical_flow.cc | 10 ++-- dali/operators/sequence/sequence_rearrange.cc | 4 +- dali/operators/sequence/sequence_rearrange.cu | 4 +- dali/operators/sequence/sequence_rearrange.h | 4 +- .../signal/decibel/to_decibels_op_cpu.cc | 8 +-- .../signal/decibel/to_decibels_op_gpu.cu | 8 +-- dali/operators/signal/fft/power_spectrum.cc | 8 +-- dali/operators/signal/fft/spectrogram.cc | 8 +-- dali/operators/signal/fft/spectrogram_gpu.cc | 6 +- dali/operators/ssd/box_encoder.cc | 8 +-- dali/operators/ssd/box_encoder.cu | 8 +-- dali/operators/ssd/box_encoder_test.cc | 8 +-- dali/operators/ssd/random_crop.cc | 18 +++--- dali/pipeline/executor/executor.cc | 14 ++--- dali/pipeline/executor/executor.h | 4 +- dali/pipeline/executor/executor_test.cc | 6 +- dali/pipeline/operator/builtin/copy.h | 6 +- .../operator/builtin/external_source.cc | 6 +- .../operator/builtin/external_source.cu | 2 +- .../operator/builtin/external_source_test.cc | 6 +- .../operator/builtin/make_contiguous.cc | 6 +- .../operator/builtin/make_contiguous.cu | 6 +- .../operator/builtin/make_contiguous.h | 4 +- dali/pipeline/operator/op_spec_test.cc | 6 +- dali/pipeline/operator/operator.cc | 6 +- dali/pipeline/operator/operator.h | 14 ++--- dali/pipeline/pipeline_test.cc | 26 ++++---- dali/pipeline/workspace/sample_workspace.cc | 8 +-- dali/pipeline/workspace/workspace.h | 16 ++--- dali/test/dali_operator_test.h | 4 +- dali/test/dali_test_bboxes.h | 14 ++--- dali/test/dali_test_matching.h | 4 +- dali/test/dali_test_resize.h | 4 +- dali/test/dali_test_single_op.h | 4 +- dali/test/plugins/dummy/dummy.cc | 4 +- dali/test/plugins/dummy/dummy.cu | 4 +- dali/test/plugins/dummy/dummy.h | 2 +- .../create_a_custom_operator.ipynb | 10 ++-- .../custom_operator/customdummy/dummy.cc | 4 +- .../custom_operator/customdummy/dummy.cu | 4 +- .../custom_operator/customdummy/dummy.h | 2 +- 198 files changed, 677 insertions(+), 677 deletions(-) diff --git a/dali/benchmark/caffe2_alexnet_bench.cc b/dali/benchmark/caffe2_alexnet_bench.cc index cd43eb4b9c3..26c6299e9da 100755 --- a/dali/benchmark/caffe2_alexnet_bench.cc +++ b/dali/benchmark/caffe2_alexnet_bench.cc @@ -127,7 +127,7 @@ BENCHMARK_DEFINE_F(C2Alexnet, Caffe2Pipe)(benchmark::State& st) { // NOLINT } } - WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); @@ -258,7 +258,7 @@ BENCHMARK_DEFINE_F(C2Alexnet, HybridPipe)(benchmark::State& st) { // NOLINT } } - // WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + // WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); diff --git a/dali/benchmark/caffe_alexnet_bench.cc b/dali/benchmark/caffe_alexnet_bench.cc index d4369e71c26..38d006322d5 100755 --- a/dali/benchmark/caffe_alexnet_bench.cc +++ b/dali/benchmark/caffe_alexnet_bench.cc @@ -128,7 +128,7 @@ BENCHMARK_DEFINE_F(Alexnet, CaffePipe)(benchmark::State& st) { // NOLINT } } - WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); @@ -259,7 +259,7 @@ BENCHMARK_DEFINE_F(Alexnet, HybridPipe)(benchmark::State& st) { // NOLINT } } - // WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + // WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); diff --git a/dali/benchmark/decoder_bench.cc b/dali/benchmark/decoder_bench.cc index c6e2294de6f..d44259f1fa7 100644 --- a/dali/benchmark/decoder_bench.cc +++ b/dali/benchmark/decoder_bench.cc @@ -81,7 +81,7 @@ class DecoderBench : public DALIBenchmark { } } - // WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + // WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + 1; st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); diff --git a/dali/benchmark/file_reader_alexnet_bench.cc b/dali/benchmark/file_reader_alexnet_bench.cc index e5989388b20..ba3aa4dc6ef 100755 --- a/dali/benchmark/file_reader_alexnet_bench.cc +++ b/dali/benchmark/file_reader_alexnet_bench.cc @@ -128,7 +128,7 @@ BENCHMARK_DEFINE_F(FileReaderAlexnet, CaffePipe)(benchmark::State& st) { // NOLI } } - WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); diff --git a/dali/benchmark/resnet50_bench.cc b/dali/benchmark/resnet50_bench.cc index 1bbf1c1ed07..12633d258d3 100755 --- a/dali/benchmark/resnet50_bench.cc +++ b/dali/benchmark/resnet50_bench.cc @@ -131,7 +131,7 @@ BENCHMARK_DEFINE_F(RN50, C2Pipe)(benchmark::State& st) { // NOLINT } } - // WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + // WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); @@ -266,7 +266,7 @@ BENCHMARK_DEFINE_F(RN50, HybridPipe)(benchmark::State& st) { // NOLINT } } - // WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + // WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); @@ -379,7 +379,7 @@ BENCHMARK_DEFINE_F(RN50, nvJPEGPipe)(benchmark::State& st) { // NOLINT } } - // WriteCHWBatch(ws.OutputRef(0), 128, 1, "img"); + // WriteCHWBatch(ws.Output(0), 128, 1, "img"); int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, benchmark::Counter::kIsRate); diff --git a/dali/benchmark/resnet50_nvjpeg_bench.cc b/dali/benchmark/resnet50_nvjpeg_bench.cc index b089be10e2f..d25ce646ef3 100755 --- a/dali/benchmark/resnet50_nvjpeg_bench.cc +++ b/dali/benchmark/resnet50_nvjpeg_bench.cc @@ -112,7 +112,7 @@ BENCHMARK_DEFINE_F(RealRN50, nvjpegPipe)(benchmark::State& st) { // NOLINT } #if DALI_DEBUG - WriteHWCBatch(ws.OutputRef(0), "img"); + WriteHWCBatch(ws.Output(0), "img"); #endif int num_batches = st.iterations() + static_cast(pipelined); st.counters["FPS"] = benchmark::Counter(batch_size*num_batches, diff --git a/dali/c_api/c_api.cc b/dali/c_api/c_api.cc index f1e0860ded7..9c8669c07db 100644 --- a/dali/c_api/c_api.cc +++ b/dali/c_api/c_api.cc @@ -318,9 +318,9 @@ void daliOutputRelease(daliPipelineHandle *pipe_handle) { int64_t daliOutputHasUniformShape(daliPipelineHandle* pipe_handle, int i) { dali::DeviceWorkspace* ws = reinterpret_cast(pipe_handle->ws); if (ws->OutputIsType(i)) { - return is_uniform(ws->OutputRef(i).shape()); + return is_uniform(ws->Output(i).shape()); } else { - return is_uniform(ws->OutputRef(i).shape()); + return is_uniform(ws->Output(i).shape()); } } @@ -328,7 +328,7 @@ template static int64_t *daliShapeAtHelper(dali::DeviceWorkspace *ws, int n, int k) { int64_t *c_shape = nullptr; std::vector shape; - const auto &out_tensor_list = ws->OutputRef(n); + const auto &out_tensor_list = ws->Output(n); if (k >= 0) { auto shape_span = out_tensor_list.tensor_shape_span(k); shape = std::vector(shape_span.begin(), shape_span.end()); @@ -366,7 +366,7 @@ int64_t* daliShapeAt(daliPipelineHandle* pipe_handle, int n) { template static dali_data_type_t daliTypeAtHelper(dali::DeviceWorkspace* ws, int n) { - const auto &out_tensor_list = ws->OutputRef(n); + const auto &out_tensor_list = ws->Output(n); auto type_id = out_tensor_list.type(); return static_cast(static_cast(type_id)); } @@ -383,7 +383,7 @@ dali_data_type_t daliTypeAt(daliPipelineHandle* pipe_handle, int n) { template static size_t daliNumTensorsHelper(dali::DeviceWorkspace* ws, int n) { - return ws->OutputRef(n).num_samples(); + return ws->Output(n).num_samples(); } size_t daliNumTensors(daliPipelineHandle* pipe_handle, int n) { @@ -397,7 +397,7 @@ size_t daliNumTensors(daliPipelineHandle* pipe_handle, int n) { template static size_t daliNumElementsHelper(dali::DeviceWorkspace* ws, int n) { - return ws->OutputRef(n)._num_elements(); + return ws->Output(n)._num_elements(); } size_t daliNumElements(daliPipelineHandle* pipe_handle, int n) { @@ -411,7 +411,7 @@ size_t daliNumElements(daliPipelineHandle* pipe_handle, int n) { template static size_t daliTensorSizeHelper(dali::DeviceWorkspace* ws, int n) { - return ws->OutputRef(n).nbytes(); + return ws->Output(n).nbytes(); } size_t daliTensorSize(daliPipelineHandle* pipe_handle, int n) { @@ -425,7 +425,7 @@ size_t daliTensorSize(daliPipelineHandle* pipe_handle, int n) { template static size_t daliMaxDimTensorsHelper(dali::DeviceWorkspace* ws, int n) { - const auto &out_tensor_list = ws->OutputRef(n); + const auto &out_tensor_list = ws->Output(n); size_t tensors_num = out_tensor_list.num_samples(); int max_num_dim = 0; for (size_t i = 0; i < tensors_num; ++i) { @@ -481,10 +481,10 @@ void daliOutputCopy(daliPipelineHandle *pipe_handle, void *dst, int output_idx, auto &type_info = dali::TypeTable::GetTypeInfo(dali::DALIDataType::DALI_UINT8); if (ws->OutputIsType(output_idx)) { - CopyToExternal(dst, dst_mem_kind, ws->OutputRef(output_idx), + CopyToExternal(dst, dst_mem_kind, ws->Output(output_idx), stream, use_copy_kernel); } else { - CopyToExternal(dst, dst_mem_kind, ws->OutputRef(output_idx), + CopyToExternal(dst, dst_mem_kind, ws->Output(output_idx), stream, use_copy_kernel); } if (sync) { @@ -506,10 +506,10 @@ void daliOutputCopySamples(daliPipelineHandle *pipe_handle, void **dsts, int out auto &type_info = dali::TypeTable::GetTypeInfo(dali::DALIDataType::DALI_UINT8); if (ws->OutputIsType(output_idx)) { - CopyToExternal(dsts, dst_mem_kind, ws->OutputRef(output_idx), + CopyToExternal(dsts, dst_mem_kind, ws->Output(output_idx), stream, use_copy_kernel); } else { - CopyToExternal(dsts, dst_mem_kind, ws->OutputRef(output_idx), + CopyToExternal(dsts, dst_mem_kind, ws->Output(output_idx), stream, use_copy_kernel); } if (sync) { diff --git a/dali/c_api/c_api_test.cc b/dali/c_api/c_api_test.cc index df83ef5c1f0..644e7974750 100644 --- a/dali/c_api/c_api_test.cc +++ b/dali/c_api/c_api_test.cc @@ -152,7 +152,7 @@ void ComparePipelinesOutputs(daliPipelineHandle &handle, Pipeline &baseline, EXPECT_EQ(daliNumTensors(&handle, output), batch_size); for (int elem = 0; elem < batch_size; elem++) { auto *shape = daliShapeAtSample(&handle, output, elem); - auto ref_shape = ws.OutputRef(output).shape()[elem]; + auto ref_shape = ws.Output(output).shape()[elem]; int D = ref_shape.size(); for (int d = 0; d < D; d++) EXPECT_EQ(shape[d], ref_shape[d]); @@ -162,7 +162,7 @@ void ComparePipelinesOutputs(daliPipelineHandle &handle, Pipeline &baseline, TensorList pipeline_output_cpu, c_api_output_cpu; // Unnecessary copy in case of CPUBackend, makes the code generic across Backends - pipeline_output_cpu.Copy(ws.OutputRef(0), cuda_stream); + pipeline_output_cpu.Copy(ws.Output(0), cuda_stream); auto num_elems = pipeline_output_cpu.shape().num_elements(); auto backend_buf = AllocBuffer(num_elems * sizeof(uint8_t), false); diff --git a/dali/operators/audio/mel_scale/mel_filter_bank.cc b/dali/operators/audio/mel_scale/mel_filter_bank.cc index 535db8b0903..74333c96f5a 100644 --- a/dali/operators/audio/mel_scale/mel_filter_bank.cc +++ b/dali/operators/audio/mel_scale/mel_filter_bank.cc @@ -69,7 +69,7 @@ template <> bool MelFilterBank::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(kNumOutputs); - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); auto in_shape = input.shape(); int nsamples = input.num_samples(); auto nthreads = ws.GetThreadPool().NumThreads(); @@ -95,8 +95,8 @@ bool MelFilterBank::SetupImpl(std::vector &output_desc, template <> void MelFilterBank::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_shape = input.shape(); auto& thread_pool = ws.GetThreadPool(); diff --git a/dali/operators/audio/mel_scale/mel_filter_bank_gpu.cc b/dali/operators/audio/mel_scale/mel_filter_bank_gpu.cc index 9082a7147ca..c7da889055f 100644 --- a/dali/operators/audio/mel_scale/mel_filter_bank_gpu.cc +++ b/dali/operators/audio/mel_scale/mel_filter_bank_gpu.cc @@ -23,7 +23,7 @@ template <> bool MelFilterBank::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(kNumOutputs); - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &in_shape = input.shape(); auto layout = input.GetLayout(); auto ndim = in_shape.sample_dim(); @@ -45,8 +45,8 @@ bool MelFilterBank::SetupImpl(std::vector &output_desc, template <> void MelFilterBank::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); const auto &in_shape = input.shape(); ctx_.gpu.stream = ws.stream(); TYPE_SWITCH(input.type(), type2id, T, MEL_FBANK_SUPPORTED_TYPES, ( diff --git a/dali/operators/audio/mfcc/mfcc.cc b/dali/operators/audio/mfcc/mfcc.cc index adb93d64aeb..4dab0b89bd0 100644 --- a/dali/operators/audio/mfcc/mfcc.cc +++ b/dali/operators/audio/mfcc/mfcc.cc @@ -115,8 +115,8 @@ bool MFCC::SetupImpl(std::vector &output_desc, const workspace_t &ws) { GetArguments(ws); output_desc.resize(kNumOutputs); - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); kernels::KernelContext ctx; auto in_shape = input.shape(); int nsamples = input.num_samples(); @@ -153,8 +153,8 @@ bool MFCC::SetupImpl(std::vector &output_desc, template <> void MFCC::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_shape = input.shape(); auto& thread_pool = ws.GetThreadPool(); diff --git a/dali/operators/audio/mfcc/mfcc.cu b/dali/operators/audio/mfcc/mfcc.cu index 5212023647a..3c4bbf095d8 100644 --- a/dali/operators/audio/mfcc/mfcc.cu +++ b/dali/operators/audio/mfcc/mfcc.cu @@ -65,7 +65,7 @@ bool MFCC::SetupImpl(std::vector &output_desc, const workspace_t &ws) { GetArguments(ws); ctx_.gpu.stream = ws.stream(); - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); auto in_shape = input.shape(); int ndim = in_shape.sample_dim(); @@ -87,11 +87,11 @@ bool MFCC::SetupImpl(std::vector &output_desc, template<> void MFCC::RunImpl(workspace_t &ws) { - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); TYPE_SWITCH(input.type(), type2id, T, MFCC_SUPPORTED_TYPES, ( using Kernel = kernels::signal::dct::Dct1DGpu; auto in_view = view(input); - auto out_view = view(ws.OutputRef(0)); + auto out_view = view(ws.Output(0)); auto lifter_view = make_tensor_gpu<1>(lifter_coeffs_.data(), {static_cast(lifter_coeffs_.size())}); kmgr_.Run(0, 0, ctx_, out_view, in_view, lifter_view); diff --git a/dali/operators/audio/mfcc/mfcc.h b/dali/operators/audio/mfcc/mfcc.h index ede41f28a21..669a49f5c38 100644 --- a/dali/operators/audio/mfcc/mfcc.h +++ b/dali/operators/audio/mfcc/mfcc.h @@ -86,7 +86,7 @@ class MFCC : public Operator { using Operator::RunImpl; void GetArguments(const workspace_t &ws) { - auto nsamples = ws.template InputRef(0).shape().size(); + auto nsamples = ws.template Input(0).shape().size(); DctArgs arg; arg.ndct = spec_.template GetArgument("n_mfcc"); DALI_ENFORCE(arg.ndct > 0, "number of MFCCs should be > 0"); diff --git a/dali/operators/audio/nonsilence_op.cc b/dali/operators/audio/nonsilence_op.cc index 9eb4d0f32c4..fc9ea4df618 100644 --- a/dali/operators/audio/nonsilence_op.cc +++ b/dali/operators/audio/nonsilence_op.cc @@ -80,7 +80,7 @@ bool NonsilenceOperatorCpu::SetupImpl(std::vector &output_desc, void NonsilenceOperatorCpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); TYPE_SWITCH(input.type(), type2id, InputType, NONSILENCE_TYPES, ( RunImplTyped(ws); ), DALI_FAIL(make_string("Unsupported input type: ", input.type()))) // NOLINT diff --git a/dali/operators/audio/nonsilence_op.h b/dali/operators/audio/nonsilence_op.h index 2f6888b9a88..45b9c689c6d 100644 --- a/dali/operators/audio/nonsilence_op.h +++ b/dali/operators/audio/nonsilence_op.h @@ -161,7 +161,7 @@ class NonsilenceOperator : public Operator { reference_max_ = true; } window_length_ = spec.GetArgument("window_length", &ws); - auto input_type = ws.template InputRef(0).type(); + auto input_type = ws.template Input(0).type(); // If input type is not floating point, there's no need for reset interval reset_interval_ = IsFloatingPoint(input_type) ? spec.GetArgument("reset_interval", &ws) : -1; @@ -207,9 +207,9 @@ class NonsilenceOperatorCpu : public NonsilenceOperator { private: template void RunImplTyped(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output_begin = ws.OutputRef(0); - auto &output_length = ws.OutputRef(1); + const auto &input = ws.template Input(0); + auto &output_begin = ws.Output(0); + auto &output_length = ws.Output(1); auto curr_batch_size = ws.GetInputBatchSize(0); auto &tp = ws.GetThreadPool(); auto in_shape = input.shape(); diff --git a/dali/operators/audio/preemphasis_filter_op.cc b/dali/operators/audio/preemphasis_filter_op.cc index c88b4d10898..fc319a4b1bc 100644 --- a/dali/operators/audio/preemphasis_filter_op.cc +++ b/dali/operators/audio/preemphasis_filter_op.cc @@ -56,8 +56,8 @@ class PreemphasisFilterCPU : public PreemphasisFilter { template void PreemphasisFilterCPU::RunImplTyped(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.Output(0); auto &tp = ws.GetThreadPool(); auto shape = input.shape(); auto nsamples = shape.num_samples(); @@ -92,7 +92,7 @@ void PreemphasisFilterCPU::RunImplTyped(workspace_t &ws) { } void PreemphasisFilterCPU::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); TYPE_SWITCH(input.type(), type2id, InputType, PREEMPH_TYPES, ( TYPE_SWITCH(output_type_, type2id, OutputType, PREEMPH_TYPES, ( RunImplTyped(ws); diff --git a/dali/operators/audio/preemphasis_filter_op.cu b/dali/operators/audio/preemphasis_filter_op.cu index 962550b0f5f..bd00c031df9 100644 --- a/dali/operators/audio/preemphasis_filter_op.cu +++ b/dali/operators/audio/preemphasis_filter_op.cu @@ -77,8 +77,8 @@ class PreemphasisFilterGPU : public PreemphasisFilter { template void PreemphasisFilterGPU::RunImplTyped(workspace_t &ws) { using SampleDesc = detail::SampleDescriptor; - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto curr_batch_size = ws.GetInputBatchSize(0); std::vector samples_cpu(curr_batch_size); @@ -104,7 +104,7 @@ void PreemphasisFilterGPU::RunImplTyped(workspace_t &ws) { } void PreemphasisFilterGPU::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); TYPE_SWITCH(input.type(), type2id, InputType, PREEMPH_TYPES, ( TYPE_SWITCH(output_type_, type2id, OutputType, PREEMPH_TYPES, ( RunImplTyped(ws); diff --git a/dali/operators/audio/preemphasis_filter_op.h b/dali/operators/audio/preemphasis_filter_op.h index 4f01d9855dd..1b8a7d81508 100644 --- a/dali/operators/audio/preemphasis_filter_op.h +++ b/dali/operators/audio/preemphasis_filter_op.h @@ -67,7 +67,7 @@ class PreemphasisFilter : public Operator { bool SetupImpl(std::vector<::dali::OutputDesc> &output_desc, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); AcquireArguments(ws); output_desc.resize(detail::kNumOutputs); auto shape = input.shape(); diff --git a/dali/operators/bbox/bb_flip.cc b/dali/operators/bbox/bb_flip.cc index 76253db5506..2925e8bd1da 100644 --- a/dali/operators/bbox/bb_flip.cc +++ b/dali/operators/bbox/bb_flip.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -40,8 +40,8 @@ system, that is 0.0-1.0)code") 0, true); void BbFlipCPU::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_view = view(input); auto out_view = view(output); auto nsamples = in_view.shape.size(); diff --git a/dali/operators/bbox/bb_flip.cu b/dali/operators/bbox/bb_flip.cu index 1ab5391f927..1285abe7645 100644 --- a/dali/operators/bbox/bb_flip.cu +++ b/dali/operators/bbox/bb_flip.cu @@ -68,10 +68,10 @@ TensorListShape<2> GetNormalizedShape(const TensorListShape<-1> &shape) { } void BbFlipGPU::RunImpl(workspace_t &ws) { - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); const auto &shape = input.shape(); auto nsamples = shape.num_samples(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); DALI_ENFORCE(IsType(input.type()), make_string("Expected input data as float; got ", input.type())); diff --git a/dali/operators/bbox/bb_flip.h b/dali/operators/bbox/bb_flip.h index 27e438bc308..bed5659c64d 100644 --- a/dali/operators/bbox/bb_flip.h +++ b/dali/operators/bbox/bb_flip.h @@ -42,7 +42,7 @@ class BbFlip : public Operator { } bool SetupImpl(std::vector &output_descs, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); DALI_ENFORCE(input.type() == DALI_FLOAT, "Bounding box in wrong format"); auto nsamples = input.shape().size(); horz_.Acquire(spec_, ws, nsamples, TensorShape<0>{}); diff --git a/dali/operators/bbox/bbox_paste.cc b/dali/operators/bbox/bbox_paste.cc index 82aee69da84..4b8588d8e21 100644 --- a/dali/operators/bbox/bbox_paste.cc +++ b/dali/operators/bbox/bbox_paste.cc @@ -56,14 +56,14 @@ canvas and ``(1,1)`` aligns it to bottom-right. template<> void BBoxPaste::RunImpl(Workspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto input_data = input.data(); DALI_ENFORCE(input.type() == DALI_FLOAT, "Bounding box in wrong format"); DALI_ENFORCE(input.size() % 4 == 0, "Bounding box tensor size must be a multiple of 4." "Got: " + std::to_string(input.size())); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.Resize(input.shape(), DALI_FLOAT); auto *output_data = output.mutable_data(); diff --git a/dali/operators/debug/dump_image.cc b/dali/operators/debug/dump_image.cc index 29740871d0f..09a3cc87de3 100644 --- a/dali/operators/debug/dump_image.cc +++ b/dali/operators/debug/dump_image.cc @@ -21,8 +21,8 @@ namespace dali { template<> void DumpImage::RunImpl(SampleWorkspace &ws) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); DALI_ENFORCE(input.ndim() == 3, make_string("Input images must have three dimensions, got input with `", diff --git a/dali/operators/debug/dump_image.cu b/dali/operators/debug/dump_image.cu index 913ce81da21..7f9381f53dc 100644 --- a/dali/operators/debug/dump_image.cu +++ b/dali/operators/debug/dump_image.cu @@ -20,8 +20,8 @@ namespace dali { template<> void DumpImage::RunImpl(DeviceWorkspace &ws) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); DALI_ENFORCE(input.shape().sample_dim() == 3, diff --git a/dali/operators/decoder/audio/audio_decoder_op.cc b/dali/operators/decoder/audio/audio_decoder_op.cc index f29080c5cd0..a624714e59c 100644 --- a/dali/operators/decoder/audio/audio_decoder_op.cc +++ b/dali/operators/decoder/audio/audio_decoder_op.cc @@ -65,7 +65,7 @@ functionality to allow for backward compatibility.)code"); // Deprecated in 1.0 bool AudioDecoderCpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); const auto batch_size = input.shape().num_samples(); GetPerSampleArgument(target_sample_rates_, "sample_rate", ws, batch_size); @@ -139,8 +139,8 @@ AudioDecoderCpu::DecodeSample(const TensorView void AudioDecoderCpu::DecodeBatch(workspace_t &ws) { - auto decoded_output = view(ws.template OutputRef(0)); - auto sample_rate_output = view(ws.template OutputRef(1)); + auto decoded_output = view(ws.template Output(0)); + auto sample_rate_output = view(ws.template Output(1)); int batch_size = decoded_output.shape.num_samples(); auto &tp = ws.GetThreadPool(); diff --git a/dali/operators/decoder/host/host_decoder.cc b/dali/operators/decoder/host/host_decoder.cc index afc37b83896..9ce0513d7a9 100644 --- a/dali/operators/decoder/host/host_decoder.cc +++ b/dali/operators/decoder/host/host_decoder.cc @@ -21,8 +21,8 @@ namespace dali { void HostDecoder::RunImpl(SampleWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto file_name = input.GetSourceInfo(); // Verify input diff --git a/dali/operators/decoder/nvjpeg/nvjpeg_decoder_decoupled_api.h b/dali/operators/decoder/nvjpeg/nvjpeg_decoder_decoupled_api.h index 267b27dd8d9..956d9bd0417 100644 --- a/dali/operators/decoder/nvjpeg/nvjpeg_decoder_decoupled_api.h +++ b/dali/operators/decoder/nvjpeg/nvjpeg_decoder_decoupled_api.h @@ -554,7 +554,7 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { #endif // NVJPEG2K_ENABLED for (int i = 0; i < curr_batch_size; i++) { - const auto &in = ws.InputRef(0)[i]; + const auto &in = ws.Input(0)[i]; const auto in_size = in.size(); thread_pool_.AddWork([this, i, &in, in_size](int tid) { auto *input_data = in.data(); @@ -683,7 +683,7 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { } void ProcessImagesCache(MixedWorkspace &ws) { - auto& output = ws.OutputRef(0); + auto& output = ws.Output(0); for (auto *sample : samples_cache_) { assert(sample); auto i = sample->sample_idx; @@ -694,12 +694,12 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { } void ProcessImagesCuda(MixedWorkspace &ws) { - auto& output = ws.OutputRef(0); + auto& output = ws.Output(0); for (auto *sample : samples_single_) { assert(sample); auto i = sample->sample_idx; auto *output_data = output.mutable_tensor(i); - const auto &in = ws.InputRef(0)[i]; + const auto &in = ws.Input(0)[i]; thread_pool_.AddWork( [this, sample, &in, output_data](int tid) { SampleWorker(sample->sample_idx, sample->file_name, in.size(), tid, @@ -783,8 +783,8 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { return; } nvjpeg2k_thread_.AddWork([this, &ws](int) { - auto &output = ws.OutputRef(0); - auto &input = ws.InputRef(0); + auto &output = ws.Output(0); + auto &input = ws.Input(0); for (auto *sample : samples_jpeg2k_) { assert(sample); auto i = sample->sample_idx; @@ -799,11 +799,11 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { } void ProcessImagesHost(MixedWorkspace &ws) { - auto& output = ws.OutputRef(0); + auto& output = ws.Output(0); for (auto *sample : samples_host_) { auto i = sample->sample_idx; auto *output_data = output.mutable_tensor(i); - const auto &in = ws.InputRef(0)[i]; + const auto &in = ws.Input(0)[i]; ImageCache::ImageShape shape = output_shape_[i].to_static<3>(); thread_pool_.AddWork( [this, sample, &in, output_data, shape](int tid) { @@ -816,7 +816,7 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { void ProcessImagesHw(MixedWorkspace &ws) { #if IS_HW_DECODER_COMPATIBLE - auto& output = ws.OutputRef(0); + auto& output = ws.Output(0); if (!samples_hw_batched_.empty()) { nvjpegJpegState_t &state = state_hw_batched_; assert(state != nullptr); @@ -839,7 +839,7 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { for (auto *sample : samples_hw_batched_) { int i = sample->sample_idx; - const auto &in = ws.InputRef(0)[i]; + const auto &in = ws.Input(0)[i]; const auto &out_shape = output_shape_.tensor_shape(i); tv[j].ShareData(const_cast &>(in)); @@ -891,7 +891,7 @@ class nvJPEGDecoder : public Operator, CachedDecoderImpl { } void ProcessImages(MixedWorkspace &ws) { - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); assert(output_shape_.num_samples() == ws.GetInputBatchSize(0)); // If fails: Incorrect number of samples in shape output.Resize(output_shape_, DALI_UINT8); diff --git a/dali/operators/generic/cast.cc b/dali/operators/generic/cast.cc index fca15be0bf1..cc9bf8137b3 100644 --- a/dali/operators/generic/cast.cc +++ b/dali/operators/generic/cast.cc @@ -30,9 +30,9 @@ void Cast::PrepareBlocks(const HostWorkspace &ws) {} template <> void Cast::RunImpl(HostWorkspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &input_shape = input.shape(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); auto num_samples = input_shape.num_samples(); diff --git a/dali/operators/generic/cast.cu b/dali/operators/generic/cast.cu index d19414ec55d..4dddca0749c 100644 --- a/dali/operators/generic/cast.cu +++ b/dali/operators/generic/cast.cu @@ -38,7 +38,7 @@ __global__ void BatchedCastKernel(const CastSampleDesc *samples, template <> void Cast::PrepareBlocks(const DeviceWorkspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &input_shape = input.shape(); std::array, 1> collapse_groups = {{{0, input_shape.sample_dim()}}}; auto collapsed_shape = collapse_dims<1>(input.shape(), collapse_groups); @@ -50,9 +50,9 @@ void Cast::PrepareBlocks(const DeviceWorkspace &ws) { template <> void Cast::RunImpl(DeviceWorkspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &input_shape = input.shape(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); auto num_samples = input_shape.num_samples(); diff --git a/dali/operators/generic/cast.h b/dali/operators/generic/cast.h index fd1d434fdfb..a464c1eb742 100644 --- a/dali/operators/generic/cast.h +++ b/dali/operators/generic/cast.h @@ -51,7 +51,7 @@ class Cast : public Operator { bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { output_desc.resize(1); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc[0].shape = input.shape(); output_desc[0].type = output_type_; PrepareBlocks(ws); diff --git a/dali/operators/generic/constant.cc b/dali/operators/generic/constant.cc index 8b7f8e53b91..d01a4f48537 100644 --- a/dali/operators/generic/constant.cc +++ b/dali/operators/generic/constant.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -99,7 +99,7 @@ void FillTensorVector( template <> void Constant::RunImpl(HostWorkspace &ws) { - auto &out = ws.OutputRef(0); + auto &out = ws.Output(0); if (output_.num_samples() == 0) { output_.set_pinned(out.is_pinned()); TYPE_SWITCH(output_type_, type2id, type, CONSTANT_OP_SUPPORTED_TYPES, diff --git a/dali/operators/generic/constant.cu b/dali/operators/generic/constant.cu index fb6faf22f1e..a8ddee76b60 100644 --- a/dali/operators/generic/constant.cu +++ b/dali/operators/generic/constant.cu @@ -99,7 +99,7 @@ void Constant::RunImpl(DeviceWorkspace &ws) { } ), (DALI_FAIL(make_string("Unsupported type: ", output_type_)))); // NOLINT } - auto &out = ws.OutputRef(0); + auto &out = ws.Output(0); out.Reset(); out.ShareData(output_); diff --git a/dali/operators/generic/erase/erase.cc b/dali/operators/generic/erase/erase.cc index 258b73657d3..725f3b62c38 100644 --- a/dali/operators/generic/erase/erase.cc +++ b/dali/operators/generic/erase/erase.cc @@ -178,7 +178,7 @@ class EraseImplCpu : public OpImplBase { template bool EraseImplCpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto layout = input.GetLayout(); auto type = input.type(); auto shape = input.shape(); @@ -199,8 +199,8 @@ bool EraseImplCpu::SetupImpl(std::vector &output_desc, template void EraseImplCpu::RunImpl(HostWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); int nsamples = input.num_samples(); auto& thread_pool = ws.GetThreadPool(); @@ -220,7 +220,7 @@ void EraseImplCpu::RunImpl(HostWorkspace &ws) { template <> bool Erase::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); auto in_shape = input.shape(); TYPE_SWITCH(input.type(), type2id, T, ERASE_SUPPORTED_TYPES, ( VALUE_SWITCH(in_shape.sample_dim(), Dims, ERASE_SUPPORTED_NDIMS, ( diff --git a/dali/operators/generic/erase/erase.cu b/dali/operators/generic/erase/erase.cu index eb7b83ead38..5adeaeda5c6 100644 --- a/dali/operators/generic/erase/erase.cu +++ b/dali/operators/generic/erase/erase.cu @@ -47,7 +47,7 @@ class EraseImplGpu : public OpImplBase { } bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto layout = input.GetLayout(); auto type = input.type(); auto shape = input.shape(); @@ -65,8 +65,8 @@ class EraseImplGpu : public OpImplBase { }; void RunImpl(workspace_t &ws) override { - const auto &input_ref = ws.template InputRef(0); - auto &output_ref = ws.template OutputRef(0); + const auto &input_ref = ws.template Input(0); + auto &output_ref = ws.template Output(0); output_ref.SetLayout(input_ref.GetLayout()); auto input = view(input_ref); auto output = view(output_ref); @@ -111,7 +111,7 @@ class EraseImplGpu : public OpImplBase { template <> bool Erase::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); auto in_shape = input.shape(); auto channel_dim = input.GetLayout().find('C'); TYPE_SWITCH(input.type(), type2id, T, ERASE_SUPPORTED_TYPES, ( diff --git a/dali/operators/generic/expand_dims.cc b/dali/operators/generic/expand_dims.cc index d36da91569d..4c88b878c78 100644 --- a/dali/operators/generic/expand_dims.cc +++ b/dali/operators/generic/expand_dims.cc @@ -96,7 +96,7 @@ bool ExpandDims::SetupImpl(std::vector &output_desc, const template void ExpandDims::GenerateSrcDims(const Workspace &ws) { - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); const auto &input_shape = in.shape(); int ndim = input_shape.sample_dim(); auto in_layout = in.GetLayout(); diff --git a/dali/operators/generic/flip.cc b/dali/operators/generic/flip.cc index 11be631d296..54e321101bf 100644 --- a/dali/operators/generic/flip.cc +++ b/dali/operators/generic/flip.cc @@ -59,8 +59,8 @@ void RunFlip(Tensor &output, const Tensor &input, template <> void Flip::RunImpl(Workspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto layout = input.GetLayout(); output.SetLayout(layout); output.Resize(input.shape(), input.type()); diff --git a/dali/operators/generic/flip.cu b/dali/operators/generic/flip.cu index cc86291c4eb..fb682f7b943 100644 --- a/dali/operators/generic/flip.cu +++ b/dali/operators/generic/flip.cu @@ -43,8 +43,8 @@ void RunKernel(TensorList &output, const TensorList &inp template <> void Flip::RunImpl(Workspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); output.Resize(input.shape(), input.type()); auto curr_batch_size = ws.GetInputBatchSize(0); diff --git a/dali/operators/generic/flip.h b/dali/operators/generic/flip.h index 1543273f529..2369975be32 100644 --- a/dali/operators/generic/flip.h +++ b/dali/operators/generic/flip.h @@ -1,4 +1,4 @@ -// Copyright (c) 2019, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -34,7 +34,7 @@ class Flip: public Operator { protected: bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { output_desc.resize(1); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); output_desc[0].type = input.type(); output_desc[0].shape = input.shape(); return true; diff --git a/dali/operators/generic/join.cc b/dali/operators/generic/join.cc index d413207aa2b..81dd6db4ffe 100644 --- a/dali/operators/generic/join.cc +++ b/dali/operators/generic/join.cc @@ -64,13 +64,13 @@ bool TensorJoin::SetupImpl( vector &outputs, const workspace_t &ws) { int njoin = this->spec_.NumRegularInput(); outputs.resize(1); - outputs[0].type = ws.template InputRef(0).type(); + outputs[0].type = ws.template Input(0).type(); auto &output_shape = outputs[0].shape; // Check that all inputs have the same type DALIDataType out_type = outputs[0].type; for (int i = 1; i < njoin; i++) { - DALIDataType type_id = ws.template InputRef(i).type(); + DALIDataType type_id = ws.template Input(i).type(); DALI_ENFORCE(type_id == out_type, make_string( "All inputs must have the same type.\nType of input #0: ", out_type, "\nType of input #", i, ": ", type_id)); @@ -97,7 +97,7 @@ void TensorJoin::SetupTyped( copy_idx_ = 0; for (int i = 0; i < ninp; i++) { - auto tlv = view(ws.template InputRef(i)); + auto tlv = view(ws.template Input(i)); if (new_axis || tlv.num_elements() > 0) { // when concatenating, we can skip empty inputs if (inputs.empty()) copy_idx_ = i; @@ -109,7 +109,7 @@ void TensorJoin::SetupTyped( // No non-empty inputs? Use the first one, even if it's empty. if (inputs.empty()) { - inputs.push_back(view(ws.template InputRef(0))); + inputs.push_back(view(ws.template Input(0))); } kernels::tensor_join::JoinedShape(output_shape, [&](int index) { @@ -126,7 +126,7 @@ void TensorJoin::GetInputLayout(const workspace_t &w int ninp = this->spec_.NumRegularInput(); for (int i = 0; i < ninp; i++) { - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); TensorLayout tl = in.GetLayout(); if (!tl.empty()) { if (!input_layout_.empty()) @@ -143,7 +143,7 @@ void TensorJoin::SetOutputLayout(const workspace_t & if (new_axis) { output_layout_ = {}; if (has_axis_name_) { - if (input_layout_.empty() && ws.template InputRef(0).shape().sample_dim() > 0) { + if (input_layout_.empty() && ws.template Input(0).shape().sample_dim() > 0) { DALI_FAIL("Specifying the new axis name with ``axis_name`` with non-scalar input requires " "a non-empty input layout."); } @@ -172,13 +172,13 @@ void TensorJoin::SetupAxis() { template void TensorJoin::RunImpl(workspace_t &ws) { - auto &out = ws.template OutputRef(0); + auto &out = ws.template Output(0); if (copy_idx_ >= 0) { // just one non-empty input - copy it to the output and return TensorListShape<> shape; if (new_axis) shape = out.shape(); - out.Copy(ws.template InputRef(copy_idx_), ws.has_stream() ? ws.stream() : 0); + out.Copy(ws.template Input(copy_idx_), ws.has_stream() ? ws.stream() : 0); if (new_axis) out.Resize(shape); out.SetLayout(output_layout_); diff --git a/dali/operators/generic/lookup_table.cc b/dali/operators/generic/lookup_table.cc index a091da05a7b..ce66d5d8365 100644 --- a/dali/operators/generic/lookup_table.cc +++ b/dali/operators/generic/lookup_table.cc @@ -41,9 +41,9 @@ void LookupValuesImpl(ThreadPool &tp, TensorVector &output, template <> void LookupTable::RunImpl(HostWorkspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &shape = input.shape(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); auto &tp = ws.GetThreadPool(); diff --git a/dali/operators/generic/lookup_table.cu b/dali/operators/generic/lookup_table.cu index 8ce07a7a9c9..1473ebf67c6 100644 --- a/dali/operators/generic/lookup_table.cu +++ b/dali/operators/generic/lookup_table.cu @@ -39,9 +39,9 @@ __global__ void LookupValuesImpl(const LutSampleDesc *samples, const kernels::Bl template<> void LookupTable::RunImpl(DeviceWorkspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &shape = input.shape(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); const auto stream = ws.stream(); diff --git a/dali/operators/generic/lookup_table.h b/dali/operators/generic/lookup_table.h index 9b79ea7a182..39d10e127c5 100644 --- a/dali/operators/generic/lookup_table.h +++ b/dali/operators/generic/lookup_table.h @@ -112,7 +112,7 @@ class LookupTable : public Operator { } output_desc.resize(1); output_desc[0].type = output_type_; - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc[0].shape = input.shape(); return true; } diff --git a/dali/operators/generic/one_hot.cc b/dali/operators/generic/one_hot.cc index 56c45b45094..41bdf06d13a 100644 --- a/dali/operators/generic/one_hot.cc +++ b/dali/operators/generic/one_hot.cc @@ -74,8 +74,8 @@ class OneHotCPU : public OneHot { }; void OneHotCPU::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); auto &tp = ws.GetThreadPool(); auto in_shape = input.shape(); auto num_samples = in_shape.num_samples(); diff --git a/dali/operators/generic/one_hot.cu b/dali/operators/generic/one_hot.cu index c967af2aeea..04f08d20817 100644 --- a/dali/operators/generic/one_hot.cu +++ b/dali/operators/generic/one_hot.cu @@ -42,7 +42,7 @@ class OneHotGPU : public OneHot { }; bool OneHotGPU::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); int num_samples = input.shape().num_samples(); if (num_samples != recent_n_samples_) { recent_n_samples_ = num_samples; @@ -55,8 +55,8 @@ bool OneHotGPU::SetupImpl(std::vector &output_desc, const workspace_ } void OneHotGPU::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); int output_sample_dim = output.shape().sample_dim(); int placement_axis = get_placement_axis(output_sample_dim); output.SetLayout(GetOutputLayout(ws, placement_axis, output_sample_dim)); @@ -69,8 +69,8 @@ void OneHotGPU::RunImpl(workspace_t &ws) { template void OneHotGPU::RunImplTyped(workspace_t &ws, int axis) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); int num_samples = input.shape().num_samples(); uint64_t max_out_vol = 1; diff --git a/dali/operators/generic/one_hot.h b/dali/operators/generic/one_hot.h index bdc47c06029..f2a2a09d552 100644 --- a/dali/operators/generic/one_hot.h +++ b/dali/operators/generic/one_hot.h @@ -80,7 +80,7 @@ class OneHot : public Operator { } bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); int input_sample_dim = input.shape().sample_dim(); int num_samples = input.shape().num_samples(); DALI_ENFORCE(-1 <= axis_ && axis_ <= input_sample_dim, @@ -114,7 +114,7 @@ class OneHot : public Operator { if (!new_axis_name_) { return {}; } - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto in_layout = input.GetLayout(); // .size method returns uint_8 which doesn't work well when 0 is printed in error message int in_layout_size = in_layout.size(); diff --git a/dali/operators/generic/pad.cc b/dali/operators/generic/pad.cc index 43635710921..8bcad7191c3 100644 --- a/dali/operators/generic/pad.cc +++ b/dali/operators/generic/pad.cc @@ -186,7 +186,7 @@ template <> bool Pad::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(1); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto in_shape = input.shape(); auto in_layout = input.GetLayout(); int ndim = in_shape.sample_dim(); @@ -219,8 +219,8 @@ bool Pad::SetupImpl(std::vector &output_desc, template <> void Pad::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); int nsamples = input.num_samples(); int ndim = input.shape().sample_dim(); diff --git a/dali/operators/generic/pad.cu b/dali/operators/generic/pad.cu index 11bef081727..62ceb2c074e 100644 --- a/dali/operators/generic/pad.cu +++ b/dali/operators/generic/pad.cu @@ -25,7 +25,7 @@ template <> bool Pad::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(1); - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); auto in_shape = input.shape(); auto in_layout = input.GetLayout(); int ndim = in_shape.sample_dim(); @@ -57,8 +57,8 @@ bool Pad::SetupImpl(std::vector &output_desc, template <> void Pad::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); int ndim = input.shape().sample_dim(); TYPE_SWITCH(input.type(), type2id, T, PAD_SUPPORTED_TYPES, ( diff --git a/dali/operators/generic/pad.h b/dali/operators/generic/pad.h index 31b512c384c..17890992eb4 100644 --- a/dali/operators/generic/pad.h +++ b/dali/operators/generic/pad.h @@ -1,4 +1,4 @@ -// Copyright (c) 2019, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -61,7 +61,7 @@ class Pad : public Operator { private: void ReadArguments(const OpSpec &spec, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto curr_batch_size = ws.GetInputBatchSize(0); auto in_shape = input.shape(); auto in_layout = input.GetLayout(); diff --git a/dali/operators/generic/permute_batch.cc b/dali/operators/generic/permute_batch.cc index dcddc05c43a..b833d2ea4d8 100644 --- a/dali/operators/generic/permute_batch.cc +++ b/dali/operators/generic/permute_batch.cc @@ -33,8 +33,8 @@ The indices must be within ``[0..batch_size)`` range. Repetitions and omissions DALI_INT_VEC, true); void PermuteBatch::RunImpl(HostWorkspace &ws) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); const auto &output_shape = output.shape(); output.SetLayout(input.GetLayout()); @@ -52,8 +52,8 @@ void PermuteBatch::RunImpl(HostWorkspace &ws) { } void PermuteBatch::RunImpl(DeviceWorkspace &ws) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); int N = indices_.size(); diff --git a/dali/operators/generic/permute_batch.h b/dali/operators/generic/permute_batch.h index 9c4a22632c9..192d0f07f7d 100644 --- a/dali/operators/generic/permute_batch.h +++ b/dali/operators/generic/permute_batch.h @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -30,7 +30,7 @@ class PermuteBatchBase : public Operator { bool SetupImpl(vector &outputs, const workspace_t &ws) override { outputs.resize(1); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); const auto &in_shape = input.shape(); outputs[0].type = input.type(); diff --git a/dali/operators/generic/reduce/mean.h b/dali/operators/generic/reduce/mean.h index 0de624dd661..1154f98a551 100644 --- a/dali/operators/generic/reduce/mean.h +++ b/dali/operators/generic/reduce/mean.h @@ -48,7 +48,7 @@ class MeanOp : public Reduce { } void RunImplImpl(workspace_t &ws) { - auto& in = ws.template InputRef(0); + auto& in = ws.template Input(0); DALIDataType input_type = in.type(); DALIDataType output_type = this->OutputType(input_type); diff --git a/dali/operators/generic/reduce/mean_square.h b/dali/operators/generic/reduce/mean_square.h index bd7fa3e4594..6399f555608 100644 --- a/dali/operators/generic/reduce/mean_square.h +++ b/dali/operators/generic/reduce/mean_square.h @@ -48,7 +48,7 @@ class MeanSquareOp : public Reduce { } void RunImplImpl(workspace_t &ws) { - auto& in = ws.template InputRef(0); + auto& in = ws.template Input(0); DALIDataType input_type = in.type(); DALIDataType output_type = this->OutputType(input_type); diff --git a/dali/operators/generic/reduce/reduce.h b/dali/operators/generic/reduce/reduce.h index 7fc5df9e794..7608fb09d74 100644 --- a/dali/operators/generic/reduce/reduce.h +++ b/dali/operators/generic/reduce/reduce.h @@ -83,7 +83,7 @@ class Reduce : public Operator, detail::AxesHelper { bool SetupImpl( std::vector &output_desc, const workspace_t &ws) override { output_desc.resize(1); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); output_desc[0].type = OutputType(input.type()); output_desc[0].shape = input.shape(); @@ -109,10 +109,10 @@ class Reduce : public Operator, detail::AxesHelper { template void RunTyped(HostWorkspace &ws) { - auto& in = ws.InputRef(0); + auto& in = ws.Input(0); auto in_view = view(in); - auto &out = ws.OutputRef(0); + auto &out = ws.Output(0); auto out_view = view(out); auto &thread_pool = ws.GetThreadPool(); @@ -140,10 +140,10 @@ class Reduce : public Operator, detail::AxesHelper { template void RunTyped(DeviceWorkspace &ws) { - auto& in = ws.InputRef(0); + auto& in = ws.Input(0); auto in_view = view(in); - auto &out = ws.OutputRef(0); + auto &out = ws.Output(0); auto out_view = view(out); using Kernel = ReductionType; @@ -184,7 +184,7 @@ class ReduceOp : public Reduce { explicit inline ReduceOp(const OpSpec &spec) : Reduce(spec) {} void RunImplImpl(workspace_t &ws) { - auto& in = ws.template InputRef(0); + auto& in = ws.template Input(0); DALIDataType input_type = in.type(); TYPE_SWITCH(input_type, type2id, DataType, REDUCE_TYPES, ( diff --git a/dali/operators/generic/reduce/reduce_with_mean_input.h b/dali/operators/generic/reduce/reduce_with_mean_input.h index 748719ecf59..68cf30e4c6b 100644 --- a/dali/operators/generic/reduce/reduce_with_mean_input.h +++ b/dali/operators/generic/reduce/reduce_with_mean_input.h @@ -49,7 +49,7 @@ class ReduceWithMeanInput : public Operator, detail::AxesHelper { bool SetupImpl( std::vector &output_desc, const workspace_t &ws) override { output_desc.resize(1); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); output_desc[0].type = DALI_FLOAT; output_desc[0].shape = input.shape(); @@ -69,7 +69,7 @@ class ReduceWithMeanInput : public Operator, detail::AxesHelper { } void RunImpl(workspace_t &ws) override { - auto& in = ws.template InputRef(0); + auto& in = ws.template Input(0); DALIDataType input_type = in.type(); DALIDataType output_type = DALI_FLOAT; @@ -84,13 +84,13 @@ class ReduceWithMeanInput : public Operator, detail::AxesHelper { template void RunTyped(HostWorkspace &ws) { - auto& in = ws.InputRef(0); + auto& in = ws.Input(0); auto in_view = view(in); - auto& mean = ws.InputRef(1); + auto& mean = ws.Input(1); auto mean_view = view(mean); - auto &out = ws.OutputRef(0); + auto &out = ws.Output(0); auto out_view = view(out); auto &thread_pool = ws.GetThreadPool(); @@ -130,13 +130,13 @@ class ReduceWithMeanInput : public Operator, detail::AxesHelper { template void RunTyped(DeviceWorkspace &ws) { - auto& in = ws.InputRef(0); + auto& in = ws.Input(0); auto in_view = view(in); - auto& mean = ws.InputRef(1); + auto& mean = ws.Input(1); auto mean_view = view(mean); - auto &out = ws.OutputRef(0); + auto &out = ws.Output(0); auto out_view = view(out); using Kernel = ReductionType; diff --git a/dali/operators/generic/reduce/root_mean_square.h b/dali/operators/generic/reduce/root_mean_square.h index 9be8381280b..f945aa1f10f 100644 --- a/dali/operators/generic/reduce/root_mean_square.h +++ b/dali/operators/generic/reduce/root_mean_square.h @@ -48,7 +48,7 @@ class RootMeanSquareOp : public Reduce } void RunImplImpl(workspace_t &ws) { - auto& in = ws.template InputRef(0); + auto& in = ws.template Input(0); DALIDataType input_type = in.type(); DALIDataType output_type = this->OutputType(input_type); diff --git a/dali/operators/generic/reduce/sum.h b/dali/operators/generic/reduce/sum.h index b0bf92372f9..6dcdc7f0bed 100644 --- a/dali/operators/generic/reduce/sum.h +++ b/dali/operators/generic/reduce/sum.h @@ -63,7 +63,7 @@ class SumOp : public Reduce { } void RunImplImpl(workspace_t &ws) { - auto& in = ws.template InputRef(0); + auto& in = ws.template Input(0); DALIDataType input_type = in.type(); DALIDataType output_type = this->OutputType(input_type); diff --git a/dali/operators/generic/reshape.cc b/dali/operators/generic/reshape.cc index 41a65d52a77..ad5446d76c8 100644 --- a/dali/operators/generic/reshape.cc +++ b/dali/operators/generic/reshape.cc @@ -261,7 +261,7 @@ void Reshape::ShapeFromInput(const TensorListLike &tl, bool relative) { template void Reshape::CalculateOutputShape(const Workspace &ws) { - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); input_shape_ = in.shape(); const int N = input_shape_.num_samples(); switch (shape_source_) { @@ -294,7 +294,7 @@ void Reshape::CalculateOutputShape(const Workspace &ws) { ShapeFromInput(ws.ArgumentInput("shape"), false); break; case ShapeSource::Input: - ShapeFromInput(ws.template InputRef(1), false); + ShapeFromInput(ws.template Input(1), false); break; case ShapeSource::None: if (!use_src_dims_) { @@ -379,15 +379,15 @@ TensorLayout Reshape::GetOutputLayout(const Workspace &ws) const { // layout was explicitly cleared return TensorLayout(); } - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); auto in_layout = in.GetLayout(); return in_layout.ndim() == output_shape_.sample_dim() ? in_layout : TensorLayout(); } template <> void Reshape::RunImpl(HostWorkspace &ws) { - auto &out = ws.OutputRef(0); - auto &in = ws.InputRef(0); + auto &out = ws.Output(0); + auto &in = ws.Input(0); TensorLayout layout = GetOutputLayout(ws); out.ShareData(in); out.Resize(output_shape_, output_type_->id()); @@ -401,8 +401,8 @@ void Reshape::RunImpl(HostWorkspace &ws) { template <> void Reshape::RunImpl(DeviceWorkspace &ws) { - auto &out = ws.OutputRef(0); - auto &in = ws.InputRef(0); + auto &out = ws.Output(0); + auto &in = ws.Input(0); TensorLayout layout = GetOutputLayout(ws); out.Reset(); out.ShareData(in); @@ -420,7 +420,7 @@ void Reshape::CheckSrcDims(const Workspace &ws) { return; } - const auto &in = ws.template InputRef(0); + const auto &in = ws.template Input(0); const auto &input_shape = in.shape(); const int ndim = input_shape.sample_dim(); for (size_t d = 0; d < src_dims_.size(); d++) { diff --git a/dali/operators/generic/reshape.h b/dali/operators/generic/reshape.h index 441ad7050b7..1b60b659c77 100644 --- a/dali/operators/generic/reshape.h +++ b/dali/operators/generic/reshape.h @@ -53,7 +53,7 @@ class Reshape : public Operator { inline void SetOutputType(const Workspace &ws) { output_type_ = output_type_arg_ != DALI_NO_TYPE ? &TypeTable::GetTypeInfo(output_type_arg_) - : &ws.template InputRef(0).type_info(); + : &ws.template Input(0).type_info(); } std::vector src_dims_; diff --git a/dali/operators/generic/roi_random_crop.cc b/dali/operators/generic/roi_random_crop.cc index dbc1a21d545..d212ce852b8 100644 --- a/dali/operators/generic/roi_random_crop.cc +++ b/dali/operators/generic/roi_random_crop.cc @@ -131,7 +131,7 @@ bool ROIRandomCropCPU::SetupImpl(std::vector &output_desc, } } } else { - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); in_shape_ = in.shape(); } @@ -175,7 +175,7 @@ bool ROIRandomCropCPU::SetupImpl(std::vector &output_desc, } void ROIRandomCropCPU::RunImpl(workspace_t &ws) { - auto &out_crop_start = ws.template OutputRef(0); + auto &out_crop_start = ws.template Output(0); auto crop_start = view(out_crop_start); int nsamples = crop_start.shape.size(); diff --git a/dali/operators/generic/shapes.h b/dali/operators/generic/shapes.h index c8449c85f92..3280fb8d545 100644 --- a/dali/operators/generic/shapes.h +++ b/dali/operators/generic/shapes.h @@ -99,14 +99,14 @@ class Shapes : public Operator { tmp_.set_pinned(true); } - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); tmp_.Resize(output.shape()); ConvertShape(tmp_, GetInputShape(ws)); output.Copy(tmp_, ws.stream()); } void RunBackend(HostWorkspace &ws) { - ConvertShape(ws.OutputRef(0), GetInputShape(ws)); + ConvertShape(ws.Output(0), GetInputShape(ws)); } static TensorListShape<1> ShapeShape(const TensorListShape<> &shape) { @@ -115,14 +115,14 @@ class Shapes : public Operator { static const TensorListShape<> &GetInputShape(const DeviceWorkspace &ws) { if (ws.InputIsType(0)) { - return ws.InputRef(0).shape(); + return ws.Input(0).shape(); } else { - return ws.InputRef(0).shape(); + return ws.Input(0).shape(); } } static auto GetInputShape(const HostWorkspace &ws) { - return ws.InputRef(0).shape(); + return ws.Input(0).shape(); } private: diff --git a/dali/operators/generic/slice/slice_attr.h b/dali/operators/generic/slice/slice_attr.h index c9f977242bd..61a8633f24c 100644 --- a/dali/operators/generic/slice/slice_attr.h +++ b/dali/operators/generic/slice/slice_attr.h @@ -234,8 +234,8 @@ class PositionalSliceAttr { if (ws.NumInput() != (spec_.GetSchema().MinNumInput() + 2)) return false; - const auto &crop_anchor = ws.template InputRef(1); - const auto &crop_shape = ws.template InputRef(2); + const auto &crop_anchor = ws.template Input(1); + const auto &crop_shape = ws.template Input(2); DALI_ENFORCE(crop_anchor.type() == crop_shape.type(), make_string("Anchor and shape should have the same type. Got: ", crop_anchor.type(), " and ", crop_shape.type())); diff --git a/dali/operators/generic/slice/slice_base.cc b/dali/operators/generic/slice/slice_base.cc index 2b4783f8282..38f2d69798f 100644 --- a/dali/operators/generic/slice/slice_base.cc +++ b/dali/operators/generic/slice/slice_base.cc @@ -48,7 +48,7 @@ class SliceBaseCpu : public OpImplBase { template bool SliceBaseCpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto in_shape = input.shape(); int nsamples = in_shape.num_samples(); auto nthreads = ws.GetThreadPool().NumThreads(); @@ -70,8 +70,8 @@ bool SliceBaseCpu::SetupImpl(std::vector void SliceBaseCpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(input.GetLayout()); int nsamples = input.num_samples(); @@ -94,7 +94,7 @@ void SliceBaseCpu::RunImpl(workspace_t template <> bool SliceBase::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto input_type = input.type(); auto ndim = input.shape().sample_dim(); diff --git a/dali/operators/generic/slice/slice_base.cu b/dali/operators/generic/slice/slice_base.cu index cee2ff6fdc9..33a85848767 100644 --- a/dali/operators/generic/slice/slice_base.cu +++ b/dali/operators/generic/slice/slice_base.cu @@ -38,7 +38,7 @@ class SliceBaseGpu : public OpImplBase { template bool SliceBaseGpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto in_shape = input.shape(); int nsamples = in_shape.num_samples(); @@ -56,8 +56,8 @@ bool SliceBaseGpu::SetupImpl(std::vector void SliceBaseGpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); auto in_view = view(input); auto out_view = view(output); @@ -70,7 +70,7 @@ void SliceBaseGpu::RunImpl(workspace_t template <> bool SliceBase::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto input_type = input.type(); auto ndim = input.shape().sample_dim(); diff --git a/dali/operators/generic/slice/slice_base.h b/dali/operators/generic/slice/slice_base.h index 619aa2c435f..88f3064d1c5 100644 --- a/dali/operators/generic/slice/slice_base.h +++ b/dali/operators/generic/slice/slice_base.h @@ -1,4 +1,4 @@ -// Copyright (c) 2019, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -55,7 +55,7 @@ class SliceBase : public Operator { void FillArgs(std::vector>& slice_args, const workspace_t &ws) { this->ProcessCroppingAttrs(ws); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto in_shape = input.shape(); int nsamples = in_shape.num_samples(); int ndim = in_shape.sample_dim(); diff --git a/dali/operators/generic/slice/subscript.cc b/dali/operators/generic/slice/subscript.cc index 023c01875fb..c7a32107398 100644 --- a/dali/operators/generic/slice/subscript.cc +++ b/dali/operators/generic/slice/subscript.cc @@ -68,8 +68,8 @@ DALI_SCHEMA(TensorSubscript) template <> template void TensorSubscript::RunTyped(HostWorkspace &ws) { - auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + auto &input = ws.template Input(0); + auto &output = ws.template Output(0); int N = input.num_samples(); using T = kernels::type_of_size; ThreadPool &tp = ws.GetThreadPool(); @@ -118,10 +118,10 @@ struct SubscriptDimCheck : public Operator { } void RunImpl(workspace_t &ws) override { - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); DALI_ENFORCE(num_subscripts_ <= in.sample_dim(), make_string("Too many indices (", num_subscripts_, ") for a ", in.sample_dim(), "-D tensor.")); - auto &out = ws.template OutputRef(0); + auto &out = ws.template Output(0); out.ShareData(in); } diff --git a/dali/operators/generic/slice/subscript.cu b/dali/operators/generic/slice/subscript.cu index ae1efe1c52d..f651a85af5c 100644 --- a/dali/operators/generic/slice/subscript.cu +++ b/dali/operators/generic/slice/subscript.cu @@ -22,8 +22,8 @@ namespace dali { template <> template void TensorSubscript::RunTyped(DeviceWorkspace &ws) { - auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + auto &input = ws.template Input(0); + auto &output = ws.template Output(0); int N = input.num_samples(); using T = kernels::type_of_size; using Kernel = kernels::SliceGPU; diff --git a/dali/operators/generic/slice/subscript.h b/dali/operators/generic/slice/subscript.h index e787e403090..8c6596df562 100644 --- a/dali/operators/generic/slice/subscript.h +++ b/dali/operators/generic/slice/subscript.h @@ -164,7 +164,7 @@ class TensorSubscript : public Operator { } bool SetupImpl(vector &outputs, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); DALI_ENFORCE(input.sample_dim() > 0, "Cannot apply an index to a scalar."); const auto &input_shape = input.shape(); outputs.resize(1); @@ -313,8 +313,8 @@ class TensorSubscript : public Operator { using Operator::RunImpl; void RunImpl(workspace_t &ws) override { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(GetOutputLayout(input.GetLayout())); VALUE_SWITCH(simplified_in_shape_.sample_dim(), ndim, (1, 2, 3, 4, 5, 6, 7, 8, 9, 10, 11, 12, 13, 14, 15, 16), diff --git a/dali/operators/generic/squeeze.cc b/dali/operators/generic/squeeze.cc index 54fae7005ae..c113eed2849 100644 --- a/dali/operators/generic/squeeze.cc +++ b/dali/operators/generic/squeeze.cc @@ -73,7 +73,7 @@ bool Squeeze::SetupImpl(std::vector &output_desc, const Wor template void Squeeze::GenerateSrcDims(const Workspace &ws) { - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); const auto &input_shape = in.shape(); const int ndim = input_shape.sample_dim(); auto in_layout = in.GetLayout(); diff --git a/dali/operators/generic/transpose/transpose.cc b/dali/operators/generic/transpose/transpose.cc index 4c901b9095d..294407f05e8 100644 --- a/dali/operators/generic/transpose/transpose.cc +++ b/dali/operators/generic/transpose/transpose.cc @@ -30,8 +30,8 @@ class TransposeCPU : public Transpose { explicit inline TransposeCPU(const OpSpec &spec) : Transpose(spec) {} void RunImpl(HostWorkspace& ws) { - const auto& input = ws.InputRef(0); - auto& output = ws.OutputRef(0); + const auto& input = ws.Input(0); + auto& output = ws.Output(0); output.SetLayout(output_layout_); auto& thread_pool = ws.GetThreadPool(); auto input_type = input.type(); diff --git a/dali/operators/generic/transpose/transpose.h b/dali/operators/generic/transpose/transpose.h index 103a83ff871..50696550723 100644 --- a/dali/operators/generic/transpose/transpose.h +++ b/dali/operators/generic/transpose/transpose.h @@ -1,4 +1,4 @@ -// Copyright (c) 2019-2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -73,7 +73,7 @@ class Transpose : public Operator { bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); SetOutputLayout(input); const auto &input_shape = input.shape(); diff --git a/dali/operators/generic/transpose/transpose_gpu.cc b/dali/operators/generic/transpose/transpose_gpu.cc index 1fa64caacf0..3cbb89d2eea 100644 --- a/dali/operators/generic/transpose/transpose_gpu.cc +++ b/dali/operators/generic/transpose/transpose_gpu.cc @@ -39,7 +39,7 @@ class TransposeGPU : public Transpose { protected: bool SetupImpl(vector &descs, const DeviceWorkspace &ws) override { Transpose::SetupImpl(descs, ws); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); kernels::KernelContext ctx; ctx.gpu.stream = ws.stream(); @@ -49,8 +49,8 @@ class TransposeGPU : public Transpose { } void RunImpl(DeviceWorkspace &ws) override { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(output_layout_); GetData(in_data_, input); diff --git a/dali/operators/geometry/affine_transforms/combine_transforms.cc b/dali/operators/geometry/affine_transforms/combine_transforms.cc index 8ff3d2539f6..3a24aac4d18 100644 --- a/dali/operators/geometry/affine_transforms/combine_transforms.cc +++ b/dali/operators/geometry/affine_transforms/combine_transforms.cc @@ -60,7 +60,7 @@ class CombineTransformsCPU : public Operator { bool SetupImpl(std::vector &output_descs, const workspace_t &ws) override { assert(ws.NumInput() > 1); - TensorListShape<> in0_shape = ws.template InputRef(0).shape(); + TensorListShape<> in0_shape = ws.template Input(0).shape(); ndim_ = in0_shape[0][0]; nsamples_ = in0_shape.size(); @@ -72,7 +72,7 @@ class CombineTransformsCPU : public Operator { "(ndim, ndim+1) representing an affine transform. Got: ", in0_shape)); for (int i = 0; i < ws.NumInput(); i++) { - const auto &shape = ws.template InputRef(i).shape(); + const auto &shape = ws.template Input(i).shape(); DALI_ENFORCE(shape == in0_shape, make_string("All input transforms are expected to have the same shape. Got: ", in0_shape, " and ", shape, " for the ", i, "-th input.")); @@ -97,14 +97,14 @@ class CombineTransformsCPU : public Operator { } constexpr int mat_dim = ndim + 1; - auto &out = ws.template OutputRef(0); + auto &out = ws.template Output(0); out.SetLayout({}); // no layout SmallVector, 64> in_views; assert(ws.NumInput() > 1); in_views.reserve(ws.NumInput()); for (int input_idx = 0; input_idx < ws.NumInput(); input_idx++) { - auto &in = ws.template InputRef(input_idx); + auto &in = ws.template Input(input_idx); in_views.push_back(view(in)); } auto out_view = view(out); diff --git a/dali/operators/geometry/affine_transforms/transform_base_op.h b/dali/operators/geometry/affine_transforms/transform_base_op.h index b6cff8f513d..d40ef8efe03 100644 --- a/dali/operators/geometry/affine_transforms/transform_base_op.h +++ b/dali/operators/geometry/affine_transforms/transform_base_op.h @@ -73,7 +73,7 @@ class TransformBaseOp : public Operator { has_input_ = ws.NumInput() > 0; auto curr_batch_size = has_input_ ? ws.GetInputBatchSize(0) : ws.GetRequestedBatchSize(0); if (has_input_) { - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); const auto &shape = input.shape(); DALI_ENFORCE(is_uniform(shape), "All matrices must have the same shape."); DALI_ENFORCE(input.type() == dtype_, @@ -112,7 +112,7 @@ class TransformBaseOp : public Operator { } constexpr int mat_dim = ndim + 1; - auto &out = ws.template OutputRef(0); + auto &out = ws.template Output(0); out.SetLayout({}); // no layout int64_t num_mats = This().IsConstantTransform() ? 1 : nsamples_; @@ -123,7 +123,7 @@ class TransformBaseOp : public Operator { auto out_view = view(out); if (has_input_) { - auto &in = ws.template InputRef(0); + auto &in = ws.template Input(0); auto in_view = view(in); for (int i = 0; i < nsamples_; i++) { int mat_idx = num_mats == 1 ? 0 : i; @@ -146,7 +146,7 @@ class TransformBaseOp : public Operator { int input_transform_ndim(const workspace_t &ws) const { assert(has_input_); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); const auto& shape = input.shape(); int ndims = shape[0][0]; assert(shape[0][1] == ndims + 1); diff --git a/dali/operators/geometry/coord_flip.cc b/dali/operators/geometry/coord_flip.cc index 5a566be8435..2e9ee69ef8f 100644 --- a/dali/operators/geometry/coord_flip.cc +++ b/dali/operators/geometry/coord_flip.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -61,8 +61,8 @@ class CoordFlipCPU : public CoordFlip { }; void CoordFlipCPU::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto &thread_pool = ws.GetThreadPool(); auto curr_batch_size = ws.GetInputBatchSize(0); diff --git a/dali/operators/geometry/coord_flip.cu b/dali/operators/geometry/coord_flip.cu index 34c80711612..d2a5a00133f 100644 --- a/dali/operators/geometry/coord_flip.cu +++ b/dali/operators/geometry/coord_flip.cu @@ -66,8 +66,8 @@ class CoordFlipGPU : public CoordFlip { }; void CoordFlipGPU::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto curr_batch_size = ws.GetInputBatchSize(0); sample_descs_.clear(); diff --git a/dali/operators/geometry/coord_flip.h b/dali/operators/geometry/coord_flip.h index c747882da0a..70397f2624e 100644 --- a/dali/operators/geometry/coord_flip.h +++ b/dali/operators/geometry/coord_flip.h @@ -39,7 +39,7 @@ class CoordFlip : public Operator { } bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); DALI_ENFORCE(input.type() == DALI_FLOAT, "Input is expected to be float"); output_desc.resize(1); diff --git a/dali/operators/geometry/coord_transform.cc b/dali/operators/geometry/coord_transform.cc index 793b510a687..9f8a01739cb 100644 --- a/dali/operators/geometry/coord_transform.cc +++ b/dali/operators/geometry/coord_transform.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -48,8 +48,8 @@ to the dynamic range of this type.)", template <> template void CoordTransform::RunTyped(HostWorkspace &ws) { - auto &in = ws.InputRef(0); - auto &out = ws.OutputRef(0); + auto &in = ws.Input(0); + auto &out = ws.Output(0); auto in_view = view(in); auto out_view = view(out); auto &tp = ws.GetThreadPool(); diff --git a/dali/operators/geometry/coord_transform.cu b/dali/operators/geometry/coord_transform.cu index d65103ecf76..ef4704bb582 100644 --- a/dali/operators/geometry/coord_transform.cu +++ b/dali/operators/geometry/coord_transform.cu @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -21,8 +21,8 @@ namespace dali { template <> template void CoordTransform::RunTyped(DeviceWorkspace &ws) { - auto &in = ws.InputRef(0); - auto &out = ws.OutputRef(0); + auto &in = ws.Input(0); + auto &out = ws.Output(0); auto in_view = view(in); auto out_view = view(out); diff --git a/dali/operators/geometry/coord_transform.h b/dali/operators/geometry/coord_transform.h index 1c4b2f5b765..33c593b8198 100644 --- a/dali/operators/geometry/coord_transform.h +++ b/dali/operators/geometry/coord_transform.h @@ -40,7 +40,7 @@ class CoordTransform : public Operator, private MTTransformAttr { protected: using Operator::spec_; bool SetupImpl(std::vector &output_descs, const workspace_t &ws) override { - auto &input = ws.template InputRef(0); // get a reference to the input tensor list + auto &input = ws.template Input(0); // get a reference to the input tensor list const auto &input_shape = input.shape(); // get a shape - use const-ref to avoid copying output_descs.resize(1); // only one output output_descs[0].type = dtype_; @@ -67,8 +67,8 @@ class CoordTransform : public Operator, private MTTransformAttr { } void RunImpl(workspace_t &ws) override { - auto &in = ws.template InputRef(0); - auto &out = ws.template OutputRef(0); + auto &in = ws.template Input(0); + auto &out = ws.template Output(0); out.SetLayout(in.GetLayout()); if (out.shape().num_elements() == 0) diff --git a/dali/operators/image/color/brightness_contrast.cc b/dali/operators/image/color/brightness_contrast.cc index 7ebabe0bdb0..1f0ce140eac 100644 --- a/dali/operators/image/color/brightness_contrast.cc +++ b/dali/operators/image/color/brightness_contrast.cc @@ -95,8 +95,8 @@ DALI_REGISTER_OPERATOR(Contrast, BrightnessContrastCpu, CPU); bool BrightnessContrastCpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { KMgrResize(num_threads_, max_batch_size_); - const auto &input = ws.template InputRef(0); - const auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + const auto &output = ws.template Output(0); output_desc.resize(1); AcquireArguments(ws); TYPE_SWITCH(input.type(), type2id, InputType, (uint8_t, int16_t, int32_t, float), ( @@ -114,8 +114,8 @@ bool BrightnessContrastCpu::SetupImpl(std::vector &output_desc, void BrightnessContrastCpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(input.GetLayout()); auto out_shape = output.shape(); auto& tp = ws.GetThreadPool(); diff --git a/dali/operators/image/color/brightness_contrast.cu b/dali/operators/image/color/brightness_contrast.cu index fb42505d024..4946321a0eb 100644 --- a/dali/operators/image/color/brightness_contrast.cu +++ b/dali/operators/image/color/brightness_contrast.cu @@ -32,8 +32,8 @@ DALI_REGISTER_OPERATOR(Contrast, BrightnessContrastGpu, GPU); bool BrightnessContrastGpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { KMgrResize(num_threads_, max_batch_size_); - const auto &input = ws.template InputRef(0); - const auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + const auto &output = ws.template Output(0); output_desc.resize(1); AcquireArguments(ws); int N = input.num_samples(); @@ -54,8 +54,8 @@ bool BrightnessContrastGpu::SetupImpl(std::vector &output_desc, void BrightnessContrastGpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(input.GetLayout()); TYPE_SWITCH(input.type(), type2id, InputType, (uint8_t, int16_t, int32_t, float), ( TYPE_SWITCH(output_type_, type2id, OutputType, (uint8_t, int16_t, int32_t, float), ( diff --git a/dali/operators/image/color/brightness_contrast.h b/dali/operators/image/color/brightness_contrast.h index 97cb8202cf7..d25289c4fb2 100644 --- a/dali/operators/image/color/brightness_contrast.h +++ b/dali/operators/image/color/brightness_contrast.h @@ -111,7 +111,7 @@ class BrightnessContrastOp : public Operator { contrast_ = std::vector(curr_batch_size, kDefaultContrast); } - input_type_ = ws.template InputRef(0).type(); + input_type_ = ws.template Input(0).type(); output_type_ = output_type_arg_ != DALI_NO_TYPE ? output_type_arg_ : input_type_; } diff --git a/dali/operators/image/color/color_space_conversion.cc b/dali/operators/image/color/color_space_conversion.cc index 7fe82ac7368..6046f72f86a 100644 --- a/dali/operators/image/color/color_space_conversion.cc +++ b/dali/operators/image/color/color_space_conversion.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2018-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -30,8 +30,8 @@ DALI_SCHEMA(ColorSpaceConversion) template <> void ColorSpaceConversion::RunImpl(HostWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); auto in_view = view(input); auto out_view = view(output); diff --git a/dali/operators/image/color/color_space_conversion.cu b/dali/operators/image/color/color_space_conversion.cu index 81d4b5beff3..2cdadd5c940 100644 --- a/dali/operators/image/color/color_space_conversion.cu +++ b/dali/operators/image/color/color_space_conversion.cu @@ -1,4 +1,4 @@ -// Copyright (c) 2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2018-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -24,8 +24,8 @@ namespace dali { template<> void ColorSpaceConversion::RunImpl(DeviceWorkspace &ws) { - const auto& input = ws.InputRef(0); - auto& output = ws.OutputRef(0); + const auto& input = ws.Input(0); + auto& output = ws.Output(0); output.SetLayout(input.GetLayout()); auto in_view = view(input); auto out_view = view(output); diff --git a/dali/operators/image/color/color_space_conversion.h b/dali/operators/image/color/color_space_conversion.h index 4007581ea30..dfe94c21b74 100644 --- a/dali/operators/image/color/color_space_conversion.h +++ b/dali/operators/image/color/color_space_conversion.h @@ -1,4 +1,4 @@ -// Copyright (c) 2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2018-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -39,7 +39,7 @@ class ColorSpaceConversion : public Operator { bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { output_desc.resize(1); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto in_sh = input.shape(); auto ndim = in_sh.sample_dim(); int nsamples = in_sh.num_samples(); diff --git a/dali/operators/image/color/color_twist.cc b/dali/operators/image/color/color_twist.cc index 0d567991369..f533126b240 100644 --- a/dali/operators/image/color/color_twist.cc +++ b/dali/operators/image/color/color_twist.cc @@ -140,7 +140,7 @@ DALI_REGISTER_OPERATOR(ColorTwist, ColorTwistCpu, CPU); bool ColorTwistCpu::SetupImpl(std::vector &output_desc, const HostWorkspace &ws) { KMgrResize(num_threads_, max_batch_size_); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc.resize(1); DetermineTransformation(ws); TYPE_SWITCH(input.type(), type2id, InputType, (uint8_t, int16_t, int32_t, float, float16), ( @@ -158,8 +158,8 @@ bool ColorTwistCpu::SetupImpl(std::vector &output_desc, const HostWo void ColorTwistCpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); auto out_shape = output.shape(); output.SetLayout(input.GetLayout()); auto &tp = ws.GetThreadPool(); diff --git a/dali/operators/image/color/color_twist.cu b/dali/operators/image/color/color_twist.cu index e4b10ed26f8..f793bdaa214 100644 --- a/dali/operators/image/color/color_twist.cu +++ b/dali/operators/image/color/color_twist.cu @@ -31,7 +31,7 @@ DALI_REGISTER_OPERATOR(ColorTwist, ColorTwistGpu, GPU); bool ColorTwistGpu::SetupImpl(std::vector &output_desc, const DeviceWorkspace &ws) { KMgrResize(num_threads_, max_batch_size_); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc.resize(1); DetermineTransformation(ws); TYPE_SWITCH(input.type(), type2id, InputType, (uint8_t, int16_t, int32_t, float), ( @@ -49,8 +49,8 @@ bool ColorTwistGpu::SetupImpl(std::vector &output_desc, const Device void ColorTwistGpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(input.GetLayout()); TYPE_SWITCH(input.type(), type2id, InputType, (uint8_t, int16_t, int32_t, float), ( TYPE_SWITCH(output_type_, type2id, OutputType, (uint8_t, int16_t, int32_t, float), ( diff --git a/dali/operators/image/color/color_twist.h b/dali/operators/image/color/color_twist.h index a938a3643cf..f97863618b0 100644 --- a/dali/operators/image/color/color_twist.h +++ b/dali/operators/image/color/color_twist.h @@ -137,7 +137,7 @@ class ColorTwistBase : public Operator { contrast_ = std::vector(curr_batch_size, 1); } - auto in_type = ws.template InputRef(0).type(); + auto in_type = ws.template Input(0).type(); output_type_ = output_type_arg_ != DALI_NO_TYPE ? output_type_arg_ : in_type; if (in_type == DALI_FLOAT16 || in_type == DALI_FLOAT || in_type == DALI_FLOAT64) { diff --git a/dali/operators/image/color/old_color_twist.cc b/dali/operators/image/color/old_color_twist.cc index 986503497dd..b6091195321 100644 --- a/dali/operators/image/color/old_color_twist.cc +++ b/dali/operators/image/color/old_color_twist.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -213,9 +213,9 @@ typedef NppStatus (*colorTwistFunc)(const Npp8u *pSrc, int nSrcStep, Npp8u *pDst template<> void OldColorTwistBase::RunImpl(DeviceWorkspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); DALI_ENFORCE(IsType(input.type()), "Color augmentations accept only uint8 tensors"); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.Resize(input.shape(), DALI_UINT8); output.SetLayout(input.GetLayout()); @@ -248,8 +248,8 @@ void OldColorTwistBase::RunImpl(DeviceWorkspace &ws) { template <> void OldColorTwistBase::RunImpl(SampleWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); const auto &input_shape = input.shape(); CheckParam(input, "Color augmentation"); diff --git a/dali/operators/image/convolution/gaussian_blur.cc b/dali/operators/image/convolution/gaussian_blur.cc index 535560b8859..aefc8e3e474 100644 --- a/dali/operators/image/convolution/gaussian_blur.cc +++ b/dali/operators/image/convolution/gaussian_blur.cc @@ -125,7 +125,7 @@ class GaussianBlurOpCpu : public OpImplBase { : spec_(spec), dim_desc_(dim_desc) {} bool SetupImpl(std::vector& output_desc, const workspace_t& ws) override { - const auto& input = ws.template InputRef(0); + const auto& input = ws.template Input(0); int nsamples = input.num_samples(); auto nthreads = ws.GetThreadPool().NumThreads(); @@ -151,8 +151,8 @@ class GaussianBlurOpCpu : public OpImplBase { } void RunImpl(workspace_t& ws) override { - const auto& input = ws.template InputRef(0); - auto& output = ws.template OutputRef(0); + const auto& input = ws.template Input(0); + auto& output = ws.template Output(0); output.SetLayout(input.GetLayout()); auto in_shape = input.shape(); auto& thread_pool = ws.GetThreadPool(); @@ -204,7 +204,7 @@ class GaussianBlurOpCpu : public OpImplBase { template <> bool GaussianBlur::SetupImpl(std::vector& output_desc, const workspace_t& ws) { - const auto& input = ws.template InputRef(0); + const auto& input = ws.template Input(0); auto layout = input.GetLayout(); auto dim_desc = ParseAndValidateDim(input.shape().sample_dim(), layout); dtype_ = dtype_ != DALI_NO_TYPE ? dtype_ : input.type(); diff --git a/dali/operators/image/convolution/gaussian_blur.cu b/dali/operators/image/convolution/gaussian_blur.cu index 6717a772bab..27dc5f1128c 100644 --- a/dali/operators/image/convolution/gaussian_blur.cu +++ b/dali/operators/image/convolution/gaussian_blur.cu @@ -73,7 +73,7 @@ extern template op_impl_uptr GetGaussianBlurGpuImpl(const OpSpec& template <> bool GaussianBlur::SetupImpl(std::vector& output_desc, const workspace_t& ws) { - const auto& input = ws.template InputRef(0); + const auto& input = ws.template Input(0); auto layout = input.GetLayout(); auto dim_desc = ParseAndValidateDim(input.shape().sample_dim(), layout); dtype_ = dtype_ != DALI_NO_TYPE ? dtype_ : input.type(); diff --git a/dali/operators/image/convolution/gaussian_blur_gpu.h b/dali/operators/image/convolution/gaussian_blur_gpu.h index 86a26242f55..c9736a872f5 100644 --- a/dali/operators/image/convolution/gaussian_blur_gpu.h +++ b/dali/operators/image/convolution/gaussian_blur_gpu.h @@ -50,7 +50,7 @@ class GaussianBlurOpGpu : public OpImplBase { bool SetupImpl(std::vector& output_desc, const workspace_t& ws) override { ctx_.gpu.stream = ws.stream(); - const auto& input = ws.template InputRef(0); + const auto& input = ws.template Input(0); auto processed_shape = input.shape(); int nsamples = processed_shape.num_samples(); // If we are sequence-like, make sure that all sequence elements are compressed to first dim @@ -83,8 +83,8 @@ class GaussianBlurOpGpu : public OpImplBase { } void RunImpl(workspace_t& ws) override { - const auto& input = ws.template InputRef(0); - auto& output = ws.template OutputRef(0); + const auto& input = ws.template Input(0); + auto& output = ws.template Output(0); output.SetLayout(input.GetLayout()); auto processed_shape = input.shape(); diff --git a/dali/operators/image/crop/bbox_crop.cc b/dali/operators/image/crop/bbox_crop.cc index 34cf670bd75..a99ee90b7a5 100644 --- a/dali/operators/image/crop/bbox_crop.cc +++ b/dali/operators/image/crop/bbox_crop.cc @@ -505,7 +505,7 @@ class RandomBBoxCropImpl : public OpImplBase { } void RunImpl(workspace_t &ws) override { - const auto &in_boxes = ws.template InputRef(0); + const auto &in_boxes = ws.template Input(0); auto in_boxes_view = view(in_boxes); auto in_boxes_shape = in_boxes_view.shape; int num_samples = in_boxes_shape.num_samples(); @@ -527,11 +527,11 @@ class RandomBBoxCropImpl : public OpImplBase { } tp.RunAll(); - auto &anchor_out = ws.template OutputRef(0); + auto &anchor_out = ws.template Output(0); anchor_out.Resize(uniform_list_shape(num_samples, {ndim}), DALI_FLOAT); auto anchor_out_view = view(anchor_out); - auto &shape_out = ws.template OutputRef(1); + auto &shape_out = ws.template Output(1); shape_out.Resize(uniform_list_shape(num_samples, {ndim}), DALI_FLOAT); auto shape_out_view = view(shape_out); @@ -551,7 +551,7 @@ class RandomBBoxCropImpl : public OpImplBase { sh[0] = sample_data_[sample_idx].prospective_crop.boxes.size(); sh[1] = ncoords; } - auto &bbox_out = ws.template OutputRef(2); + auto &bbox_out = ws.template Output(2); bbox_out.Resize(bbox_out_shape, DALI_FLOAT); auto bbox_out_view = view(bbox_out); for (int sample_idx = 0; sample_idx < num_samples; sample_idx++) { @@ -562,10 +562,10 @@ class RandomBBoxCropImpl : public OpImplBase { int next_out_idx = 3; if (has_labels_) { - const auto &labels_in = ws.template InputRef(1); + const auto &labels_in = ws.template Input(1); auto labels_in_view = view(labels_in); - auto &labels_out = ws.template OutputRef(next_out_idx++); + auto &labels_out = ws.template Output(next_out_idx++); TensorListShape<> labels_out_shape = labels_in.shape(); for (int sample_idx = 0; sample_idx < num_samples; sample_idx++) { auto sh = labels_out_shape.tensor_shape_span(sample_idx); @@ -588,7 +588,7 @@ class RandomBBoxCropImpl : public OpImplBase { } if (output_bbox_indices_) { - auto &bbox_indices_out = ws.template OutputRef(next_out_idx++); + auto &bbox_indices_out = ws.template Output(next_out_idx++); TensorListShape<> bbox_indices_out_shape; bbox_indices_out_shape.resize(num_samples, 1); for (int sample_idx = 0; sample_idx < num_samples; sample_idx++) { @@ -901,7 +901,7 @@ RandomBBoxCrop::RandomBBoxCrop(const OpSpec &spec) template <> bool RandomBBoxCrop::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &boxes = ws.template InputRef(0); + const auto &boxes = ws.template Input(0); auto tl_shape = boxes.shape(); DALI_ENFORCE(tl_shape.sample_dim() == 2, make_string( "Unexpected number of dimensions for bounding boxes input: ", tl_shape.sample_dim())); diff --git a/dali/operators/image/crop/crop.h b/dali/operators/image/crop/crop.h index 888eecdc384..646e2f58f1a 100644 --- a/dali/operators/image/crop/crop.h +++ b/dali/operators/image/crop/crop.h @@ -1,4 +1,4 @@ -// Copyright (c) 2019, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -38,7 +38,7 @@ class Crop : public SliceBase { protected: void ProcessCroppingAttrs(const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); const TensorLayout in_layout = input.GetLayout(); DALI_ENFORCE(in_layout.ndim() == input.shape().sample_dim()); DALI_ENFORCE(ImageLayoutInfo::HasChannel(in_layout) && diff --git a/dali/operators/image/crop/crop_mirror_normalize.cc b/dali/operators/image/crop/crop_mirror_normalize.cc index f119ab51841..21c20eb344a 100644 --- a/dali/operators/image/crop/crop_mirror_normalize.cc +++ b/dali/operators/image/crop/crop_mirror_normalize.cc @@ -75,7 +75,7 @@ bool CropMirrorNormalize::SetupImpl(std::vector &output_ const HostWorkspace &ws) { output_desc.resize(1); SetupCommonImpl(ws); - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); auto in_shape = input.shape(); int ndim = in_shape.sample_dim(); int nsamples = in_shape.size(); @@ -102,8 +102,8 @@ bool CropMirrorNormalize::SetupImpl(std::vector &output_ template <> void CropMirrorNormalize::RunImpl(HostWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(output_layout_); auto in_shape = input.shape(); auto out_shape = output.shape(); diff --git a/dali/operators/image/crop/crop_mirror_normalize.cu b/dali/operators/image/crop/crop_mirror_normalize.cu index 058ffc0f49a..041d5879020 100644 --- a/dali/operators/image/crop/crop_mirror_normalize.cu +++ b/dali/operators/image/crop/crop_mirror_normalize.cu @@ -27,7 +27,7 @@ bool CropMirrorNormalize::SetupImpl(std::vector &output_ auto curr_batch_size = ws.GetInputBatchSize(0); output_desc.resize(1); SetupCommonImpl(ws); - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); int ndim = input.shape().sample_dim(); TYPE_SWITCH(input_type_, type2id, InputType, CMN_IN_TYPES, ( TYPE_SWITCH(output_type_, type2id, OutputType, CMN_OUT_TYPES, ( @@ -52,8 +52,8 @@ bool CropMirrorNormalize::SetupImpl(std::vector &output_ template<> void CropMirrorNormalize::RunImpl(DeviceWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(output_layout_); int ndim = input.shape().sample_dim(); TYPE_SWITCH(input_type_, type2id, InputType, CMN_IN_TYPES, ( diff --git a/dali/operators/image/crop/crop_mirror_normalize.h b/dali/operators/image/crop/crop_mirror_normalize.h index 9e0ea47557d..6e59fff04fc 100755 --- a/dali/operators/image/crop/crop_mirror_normalize.h +++ b/dali/operators/image/crop/crop_mirror_normalize.h @@ -161,7 +161,7 @@ class CropMirrorNormalize : public Operator { } void SetupCommonImpl(const workspace_t &ws) { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); input_type_ = input.type(); assert(output_type_ != DALI_NO_TYPE); diff --git a/dali/operators/image/distortion/jpeg_compression_distortion_op.h b/dali/operators/image/distortion/jpeg_compression_distortion_op.h index b8145b06d15..ae6e9a6125c 100644 --- a/dali/operators/image/distortion/jpeg_compression_distortion_op.h +++ b/dali/operators/image/distortion/jpeg_compression_distortion_op.h @@ -1,4 +1,4 @@ -// Copyright (c) 2021, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -39,7 +39,7 @@ class JpegCompressionDistortion : public Operator { } bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc.resize(1); const auto &in_sh = input.shape(); assert(in_sh.sample_dim() == 3); // should be check by the layout diff --git a/dali/operators/image/distortion/jpeg_compression_distortion_op_cpu.cc b/dali/operators/image/distortion/jpeg_compression_distortion_op_cpu.cc index 00808b61da4..2b659b910b2 100644 --- a/dali/operators/image/distortion/jpeg_compression_distortion_op_cpu.cc +++ b/dali/operators/image/distortion/jpeg_compression_distortion_op_cpu.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2021, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -57,8 +57,8 @@ class JpegCompressionDistortionCPU : public JpegCompressionDistortion &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_shape = input.shape(); int nsamples = input.num_samples(); auto& thread_pool = ws.GetThreadPool(); diff --git a/dali/operators/image/distortion/jpeg_compression_distortion_op_gpu.cu b/dali/operators/image/distortion/jpeg_compression_distortion_op_gpu.cu index 91f3d90c453..0501b8ef9da 100644 --- a/dali/operators/image/distortion/jpeg_compression_distortion_op_gpu.cu +++ b/dali/operators/image/distortion/jpeg_compression_distortion_op_gpu.cu @@ -1,4 +1,4 @@ -// Copyright (c) 2021, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -39,8 +39,8 @@ class JpegCompressionDistortionGPU : public JpegCompressionDistortion &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_view = view(input); auto out_view = view(output); int nsamples = in_view.shape.size(); diff --git a/dali/operators/image/mask/grid_mask.cc b/dali/operators/image/mask/grid_mask.cc index 0facd5321b6..a505e781e95 100644 --- a/dali/operators/image/mask/grid_mask.cc +++ b/dali/operators/image/mask/grid_mask.cc @@ -47,8 +47,8 @@ width of black squares plus the spacing between them.)code", bool GridMaskCpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); - const auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + const auto &output = ws.template Output(0); output_desc.resize(1); GetArguments(ws); output_desc[0] = {input.shape(), input.type()}; @@ -64,8 +64,8 @@ bool GridMaskCpu::SetupImpl(std::vector &output_desc, void GridMaskCpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(input.GetLayout()); auto out_shape = output.shape(); auto& tp = ws.GetThreadPool(); diff --git a/dali/operators/image/mask/grid_mask.cu b/dali/operators/image/mask/grid_mask.cu index 1e69f1e642d..844ed62a6ca 100644 --- a/dali/operators/image/mask/grid_mask.cu +++ b/dali/operators/image/mask/grid_mask.cu @@ -23,8 +23,8 @@ namespace dali { bool GridMaskGpu::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.template InputRef(0); - const auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + const auto &output = ws.template Output(0); output_desc.resize(1); GetArguments(ws); output_desc[0] = {input.shape(), input.type()}; @@ -42,8 +42,8 @@ bool GridMaskGpu::SetupImpl(std::vector &output_desc, } void GridMaskGpu::RunImpl(workspace_t &ws) { - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(input.GetLayout()); TYPE_SWITCH(input.type(), type2id, Type, GRID_MASK_TYPES, ( { diff --git a/dali/operators/image/paste/multipaste.cc b/dali/operators/image/paste/multipaste.cc index a75f18dc966..aee713727fe 100644 --- a/dali/operators/image/paste/multipaste.cc +++ b/dali/operators/image/paste/multipaste.cc @@ -68,8 +68,8 @@ void MultiPasteCPU::SetupTyped(const workspace_t & /*ws*/, template void MultiPasteCPU::RunTyped(workspace_t &ws) { - const auto &images = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &images = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(images.GetLayout()); auto out_shape = output.shape(); diff --git a/dali/operators/image/paste/multipaste.cu b/dali/operators/image/paste/multipaste.cu index 64e2eeaa9e7..aea5171782d 100644 --- a/dali/operators/image/paste/multipaste.cu +++ b/dali/operators/image/paste/multipaste.cu @@ -55,7 +55,7 @@ void MultiPasteGPU::InitSamples(const TensorListShape<> &out_shape) { template void MultiPasteGPU::SetupTyped(const workspace_t &ws, const TensorListShape<> &out_shape) { - const auto &images = ws.template InputRef(0); + const auto &images = ws.template Input(0); const auto &in = view(images); using Kernel = kernels::PasteGPU; kernels::KernelContext ctx; @@ -68,8 +68,8 @@ void MultiPasteGPU::SetupTyped(const workspace_t &ws, template void MultiPasteGPU::RunTyped(workspace_t &ws) { - const auto &images = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &images = ws.template Input(0); + auto &output = ws.template Output(0); output.SetLayout(images.GetLayout()); auto out_shape = output.shape(); diff --git a/dali/operators/image/paste/multipaste.h b/dali/operators/image/paste/multipaste.h index 91af4f0145f..4ab4b4cbc9a 100644 --- a/dali/operators/image/paste/multipaste.h +++ b/dali/operators/image/paste/multipaste.h @@ -99,7 +99,7 @@ class MultiPasteOp : public Operator { } void AcquireArguments(const OpSpec &spec, const workspace_t &ws) { - const auto &images = ws.template InputRef(0); + const auto &images = ws.template Input(0); auto curr_batch_size = ws.GetRequestedBatchSize(0); if (curr_batch_size == 0) @@ -117,7 +117,7 @@ class MultiPasteOp : public Operator { if (shapes_.IsDefined()) { shapes_.Acquire(spec, ws, curr_batch_size, false); } - input_type_ = ws.template InputRef(0).type(); + input_type_ = ws.template Input(0).type(); output_type_ = output_type_arg_ != DALI_NO_TYPE ? output_type_arg_ @@ -198,7 +198,7 @@ class MultiPasteOp : public Operator { using Operator::RunImpl; void RunImpl(workspace_t &ws) override { - const auto input_type_id = ws.template InputRef(0).type(); + const auto input_type_id = ws.template Input(0).type(); TYPE_SWITCH(input_type_id, type2id, InputType, (MULTIPASTE_INPUT_TYPES), ( TYPE_SWITCH(output_type_, type2id, OutputType, (MULTIPASTE_OUTPUT_TYPES), ( This().template RunTyped(ws); @@ -208,7 +208,7 @@ class MultiPasteOp : public Operator { bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { - const auto &images = ws.template InputRef(0); + const auto &images = ws.template Input(0); int ndim = images.sample_dim(); DALI_ENFORCE(ndim == 3, "MultiPaste supports only 2D data with channels (HWC)"); diff --git a/dali/operators/image/paste/paste.cu b/dali/operators/image/paste/paste.cu index e420d0eb386..6eb3d646b91 100644 --- a/dali/operators/image/paste/paste.cu +++ b/dali/operators/image/paste/paste.cu @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -124,8 +124,8 @@ void Paste::SetupSharedSampleParams(DeviceWorkspace &ws) { template<> void Paste::SetupSampleParams(DeviceWorkspace &ws) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); auto curr_batch_size = ws.GetInputBatchSize(0); std::vector> output_shape(curr_batch_size); diff --git a/dali/operators/image/peek_shape/peek_image_shape.h b/dali/operators/image/peek_shape/peek_image_shape.h index 071f37b9fe3..7b15129182c 100644 --- a/dali/operators/image/peek_shape/peek_image_shape.h +++ b/dali/operators/image/peek_shape/peek_image_shape.h @@ -53,7 +53,7 @@ class PeekImageShape : public Operator { protected: bool SetupImpl(std::vector &output_desc, const HostWorkspace &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); size_t batch_size = input.num_samples(); output_desc.resize(1); output_desc[0].shape = uniform_list_shape<1>(batch_size, { 3 }); @@ -71,8 +71,8 @@ class PeekImageShape : public Operator { void RunImpl(HostWorkspace &ws) override { auto &thread_pool = ws.GetThreadPool(); - const auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + const auto &input = ws.template Input(0); + auto &output = ws.template Output(0); size_t batch_size = input.num_samples(); for (size_t sample_id = 0; sample_id < batch_size; ++sample_id) { diff --git a/dali/operators/image/remap/displacement_filter_impl_cpu.h b/dali/operators/image/remap/displacement_filter_impl_cpu.h index 262d6de0ef0..9625801e6c6 100644 --- a/dali/operators/image/remap/displacement_filter_impl_cpu.h +++ b/dali/operators/image/remap/displacement_filter_impl_cpu.h @@ -104,8 +104,8 @@ class DisplacementFilter } void RunSample(HostWorkspace &ws, int sample_idx, int thread_idx) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); PrepareDisplacement(ws, sample_idx, thread_idx); @@ -149,7 +149,7 @@ class DisplacementFilter } bool SetupImpl(std::vector &output_desc, const HostWorkspace &ws) override { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); output_desc.resize(1); output_desc[0].shape = input.shape(); output_desc[0].type = input.type(); @@ -157,9 +157,9 @@ class DisplacementFilter } void RunImpl(HostWorkspace &ws) override { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &shape = input.shape(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); if (has_mask_) { diff --git a/dali/operators/image/remap/displacement_filter_impl_gpu.cuh b/dali/operators/image/remap/displacement_filter_impl_gpu.cuh index 20ff4f93d14..7064e327b23 100644 --- a/dali/operators/image/remap/displacement_filter_impl_gpu.cuh +++ b/dali/operators/image/remap/displacement_filter_impl_gpu.cuh @@ -246,7 +246,7 @@ class DisplacementFilter &output_desc, const DeviceWorkspace &ws) override { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); output_desc.resize(1); output_desc[0].shape = input.shape(); output_desc[0].type = input.type(); @@ -254,8 +254,8 @@ class DisplacementFilter(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); if (IsType(input.type())) { @@ -312,9 +312,9 @@ class DisplacementFilter bool BatchedGPUKernel(DeviceWorkspace &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); const auto &shape = input.shape(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); auto stream = ws.stream(); const auto num_samples = shape.num_samples(); diff --git a/dali/operators/image/remap/rotate_params.h b/dali/operators/image/remap/rotate_params.h index 2e2a12fa056..48ff5fdb4c4 100644 --- a/dali/operators/image/remap/rotate_params.h +++ b/dali/operators/image/remap/rotate_params.h @@ -116,7 +116,7 @@ class RotateParamProvider using Base::out_sizes_; void SetParams() override { - input_shape_ = convert_dim(ws_->template InputRef(0).shape()); + input_shape_ = convert_dim(ws_->template Input(0).shape()); Collect(angles_, "angle", true); // For 2D, assume positive CCW rotation when (0,0) denotes top-left corner. diff --git a/dali/operators/image/remap/warp.h b/dali/operators/image/remap/warp.h index fb59aea6817..37f05d18d4a 100644 --- a/dali/operators/image/remap/warp.h +++ b/dali/operators/image/remap/warp.h @@ -84,7 +84,7 @@ class WarpOpImpl : public OpImplInterface { void Setup(TensorListShape<> &shape, const Workspace &ws) override { param_provider_->SetContext(Spec(), ws); - input_ = view(ws.template InputRef(0)); + input_ = view(ws.template Input(0)); param_provider_->Setup(); SetupBackend(shape, ws); @@ -149,8 +149,8 @@ class WarpOpImpl : public OpImplInterface { void RunBackend(HostWorkspace &ws) { param_provider_->SetContext(Spec(), ws); - auto output = view(ws.template OutputRef(0)); - input_ = view(ws.template InputRef(0)); + auto output = view(ws.template Output(0)); + input_ = view(ws.template Input(0)); ThreadPool &pool = ws.GetThreadPool(); auto interp_types = param_provider_->InterpTypes(); @@ -175,8 +175,8 @@ class WarpOpImpl : public OpImplInterface { void RunBackend(DeviceWorkspace &ws) { param_provider_->SetContext(Spec(), ws); - auto output = view(ws.template OutputRef(0)); - input_ = view(ws.template InputRef(0)); + auto output = view(ws.template Output(0)); + input_ = view(ws.template Input(0)); auto context = GetContext(ws); kmgr_.Run( 0, 0, context, @@ -294,7 +294,7 @@ class Warp : public Operator { void SetupWarp(TensorListShape<> &out_shape, DALIDataType &out_type, const Workspace &ws) { - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); input_shape_ = input.shape(); input_type_ = input.type(); output_type_ = output_type_arg_ == DALI_NO_TYPE ? input_type_ : output_type_arg_; @@ -328,8 +328,8 @@ class Warp : public Operator { void RunImpl(Workspace &ws) override { assert(impl_); impl_->Run(ws); - auto &out = ws.template OutputRef(0); - auto &in = ws.template InputRef(0); + auto &out = ws.template Output(0); + auto &in = ws.template Input(0); out.SetLayout(in.GetLayout()); } diff --git a/dali/operators/image/remap/warp_affine_params.h b/dali/operators/image/remap/warp_affine_params.h index f5aefdc1709..cb6baeae248 100644 --- a/dali/operators/image/remap/warp_affine_params.h +++ b/dali/operators/image/remap/warp_affine_params.h @@ -52,9 +52,9 @@ class WarpAffineParamProvider bool invert = !spec_->template GetArgument("inverse_map"); if (spec_->NumRegularInput() >= 2) { if (ws_->template InputIsType(1)) { - UseInputAsParams(ws_->template InputRef(1), invert); + UseInputAsParams(ws_->template Input(1), invert); } else { - UseInputAsParams(ws_->template InputRef(1), invert); + UseInputAsParams(ws_->template Input(1), invert); } } else if (spec_->HasTensorArgument("matrix")) { UseInputAsParams(ws_->ArgumentInput("matrix"), invert); diff --git a/dali/operators/image/remap/warp_param_provider.h b/dali/operators/image/remap/warp_param_provider.h index a35d08ff676..abf9a61b292 100644 --- a/dali/operators/image/remap/warp_param_provider.h +++ b/dali/operators/image/remap/warp_param_provider.h @@ -205,7 +205,7 @@ class WarpParamProvider : public InterpTypeProvider, public BorderTypeProvider(0).shape().num_samples(); + return ws.template Input(0).shape().num_samples(); } virtual void ResetParams() { @@ -284,7 +284,7 @@ class WarpParamProvider : public InterpTypeProvider, public BorderTypeProvidertemplate InputRef(0).shape(); + decltype(auto) input_shape = ws_->template Input(0).shape(); const int N = input_shape.num_samples(); out_sizes_.resize(N); SpatialShape scalar_size; diff --git a/dali/operators/image/resize/random_resized_crop.cc b/dali/operators/image/resize/random_resized_crop.cc index 1669ad018af..f2f913a9ba2 100644 --- a/dali/operators/image/resize/random_resized_crop.cc +++ b/dali/operators/image/resize/random_resized_crop.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -44,8 +44,8 @@ void RandomResizedCrop::BackendInit() { template<> void RandomResizedCrop::RunImpl(HostWorkspace &ws) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); RunResize(ws, output, input); output.SetLayout(input.GetLayout()); diff --git a/dali/operators/image/resize/random_resized_crop.cu b/dali/operators/image/resize/random_resized_crop.cu index 49049937417..8f5a9da7da6 100644 --- a/dali/operators/image/resize/random_resized_crop.cu +++ b/dali/operators/image/resize/random_resized_crop.cu @@ -30,8 +30,8 @@ void RandomResizedCrop::BackendInit() { template<> void RandomResizedCrop::RunImpl(DeviceWorkspace &ws) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); RunResize(ws, output, input); output.SetLayout(input.GetLayout()); } diff --git a/dali/operators/image/resize/random_resized_crop.h b/dali/operators/image/resize/random_resized_crop.h index 0566d0e08d9..4f805c7737f 100644 --- a/dali/operators/image/resize/random_resized_crop.h +++ b/dali/operators/image/resize/random_resized_crop.h @@ -53,7 +53,7 @@ class RandomResizedCrop : public Operator bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { auto curr_batch_size = ws.GetInputBatchSize(0); InitParams(curr_batch_size); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); const auto &in_shape = input.shape(); DALIDataType in_type = input.type(); diff --git a/dali/operators/image/resize/resize.cc b/dali/operators/image/resize/resize.cc index 95a506d44c5..01a2e8db201 100755 --- a/dali/operators/image/resize/resize.cc +++ b/dali/operators/image/resize/resize.cc @@ -52,15 +52,15 @@ void Resize::InitializeBackend() { template <> void Resize::RunImpl(HostWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); RunResize(ws, output, input); output.SetLayout(input.GetLayout()); if (save_attrs_) { const auto &input_shape = input.shape(); - auto &attr_out = ws.OutputRef(1); + auto &attr_out = ws.Output(1); const auto &attr_shape = attr_out.shape(); assert(attr_shape.num_samples() == input_shape.num_samples() && attr_shape.sample_dim() == 1 && @@ -82,14 +82,14 @@ void Resize::InitializeBackend() { template<> void Resize::RunImpl(DeviceWorkspace &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); RunResize(ws, output, input); output.SetLayout(input.GetLayout()); if (save_attrs_) { - auto &attr_out = ws.OutputRef(1); + auto &attr_out = ws.Output(1); const auto &attr_shape = attr_out.shape(); assert(attr_shape.num_samples() == input.shape().num_samples() && attr_shape.sample_dim() == 1 && diff --git a/dali/operators/image/resize/resize.h b/dali/operators/image/resize/resize.h index 099eccd0e45..a0469bab092 100644 --- a/dali/operators/image/resize/resize.h +++ b/dali/operators/image/resize/resize.h @@ -95,7 +95,7 @@ template bool Resize::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(save_attrs_ ? 2 : 1); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); const auto &in_shape = input.shape(); auto in_type = input.type(); diff --git a/dali/operators/image/resize/resize_crop_mirror.h b/dali/operators/image/resize/resize_crop_mirror.h index 04ff4aade6d..4defed13a95 100755 --- a/dali/operators/image/resize/resize_crop_mirror.h +++ b/dali/operators/image/resize/resize_crop_mirror.h @@ -164,10 +164,10 @@ class ResizeCropMirrorAttr : protected CropAttr { * @return const vector One matching shape for all inputs */ virtual const std::vector CheckShapes(const SampleWorkspace *ws) { - const auto &input = ws->InputRef(0); + const auto &input = ws->Input(0); // enforce that all shapes match for (int i = 1; i < ws->NumInput(); ++i) { - DALI_ENFORCE(input.SameShape(ws->InputRef(i))); + DALI_ENFORCE(input.SameShape(ws->Input(i))); } DALI_ENFORCE(input.ndim() == 3, "Operator expects 3-dimensional image input."); @@ -230,8 +230,8 @@ class ResizeCropMirror : public Operator, protected ResizeCropMirror } inline void RunResizeImpl(SampleWorkspace &ws, resizeCropMirroHost func) { - auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + auto &input = ws.Input(0); + auto &output = ws.Output(0); CheckParam(input, "ResizeCropMirror"); const TransformMeta &meta = per_thread_meta_[ws.thread_idx()]; diff --git a/dali/operators/math/expressions/arithmetic.cc b/dali/operators/math/expressions/arithmetic.cc index c19da959f46..f1cb77bce3c 100644 --- a/dali/operators/math/expressions/arithmetic.cc +++ b/dali/operators/math/expressions/arithmetic.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2019, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -24,7 +24,7 @@ void ArithmeticGenericOp::RunImpl(HostWorkspace &ws) { PrepareTilesForTasks(tiles_per_task_, exec_order_, tile_cover_, ws, constant_storage_, spec_); auto &pool = ws.GetThreadPool(); - ws.OutputRef(0).SetLayout(result_layout_); + ws.Output(0).SetLayout(result_layout_); for (size_t task_idx = 0; task_idx < tile_range_.size(); task_idx++) { pool.AddWork([this, task_idx](int thread_idx) { auto range = tile_range_[task_idx]; diff --git a/dali/operators/math/expressions/arithmetic.cu b/dali/operators/math/expressions/arithmetic.cu index 762d489fa0f..d4124b515c6 100644 --- a/dali/operators/math/expressions/arithmetic.cu +++ b/dali/operators/math/expressions/arithmetic.cu @@ -1,4 +1,4 @@ -// Copyright (c) 2019, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -21,7 +21,7 @@ template <> void ArithmeticGenericOp::RunImpl(DeviceWorkspace &ws) { PrepareTilesForTasks(tiles_per_task_, exec_order_, tile_cover_, ws, constant_storage_, spec_); - ws.OutputRef(0).SetLayout(result_layout_); + ws.Output(0).SetLayout(result_layout_); assert(tile_range_.size() == 1 && "Expected to cover whole GPU execution by 1 task"); for (size_t i = 0; i < exec_order_.size(); i++) { // call impl for whole batch diff --git a/dali/operators/math/expressions/arithmetic.h b/dali/operators/math/expressions/arithmetic.h index b34822967ed..5a352efcca3 100644 --- a/dali/operators/math/expressions/arithmetic.h +++ b/dali/operators/math/expressions/arithmetic.h @@ -75,7 +75,7 @@ DLL_PUBLIC TensorLayout GetCommonLayout(ExprNode &expr, const workspace_t(expr); - return ws.template InputRef(e.GetInputIndex()).GetLayout(); + return ws.template Input(e.GetInputIndex()).GetLayout(); } if (expr.GetSubexpressionCount() == 0) { return ""; @@ -110,7 +110,7 @@ DLL_PUBLIC DALIDataType PropagateTypes(ExprNode &expr, const workspace_t(expr); - expr.SetTypeId(ws.template InputRef(e.GetInputIndex()).type()); + expr.SetTypeId(ws.template Input(e.GetInputIndex()).type()); return expr.GetTypeId(); } auto &func = dynamic_cast(expr); @@ -192,7 +192,7 @@ DLL_PUBLIC inline const TensorListShape<> &PropagateShapes(ExprNode &expr, } if (expr.GetNodeType() == NodeType::Tensor) { auto &e = dynamic_cast(expr); - expr.SetShape(ws.template InputRef(e.GetInputIndex()).shape()); + expr.SetShape(ws.template Input(e.GetInputIndex()).shape()); return expr.GetShape(); } auto &func = dynamic_cast(expr); diff --git a/dali/operators/math/expressions/arithmetic_test.cc b/dali/operators/math/expressions/arithmetic_test.cc index 0c69cdbe378..572b0ed8dcf 100644 --- a/dali/operators/math/expressions/arithmetic_test.cc +++ b/dali/operators/math/expressions/arithmetic_test.cc @@ -226,7 +226,7 @@ class BinaryArithmeticOpsTest vector result_cpu(shape.num_elements()); auto *target_ptr = result_cpu.data(); for (int i = 0; i < shape.num_samples(); i++) { - auto *result = ws.OutputRef(0).template tensor(i); + auto *result = ws.Output(0).template tensor(i); MemCopy(target_ptr, result, shape[i].num_elements() * sizeof(T)); target_ptr += shape[i].num_elements(); } @@ -343,8 +343,8 @@ TEST(ArithmeticOpsTest, GenericPipeline) { vector result2_cpu(tensor_elements); for (int sample_id = 0; sample_id < batch_size; sample_id++) { const auto *data = batch.tensor(sample_id); - auto *result = ws.OutputRef(0).tensor(sample_id); - auto *result2 = ws.OutputRef(1).tensor(sample_id); + auto *result = ws.Output(0).tensor(sample_id); + auto *result2 = ws.Output(1).tensor(sample_id); MemCopy(result2_cpu.data(), result2, tensor_elements * sizeof(int)); CUDA_CALL(cudaStreamSynchronize(0)); @@ -396,16 +396,16 @@ TEST(ArithmeticOpsTest, FdivPipeline) { pipe.RunGPU(); DeviceWorkspace ws; pipe.Outputs(&ws); - ASSERT_EQ(ws.OutputRef(0).type(), DALI_FLOAT); - ASSERT_EQ(ws.OutputRef(1).type(), DALI_FLOAT); + ASSERT_EQ(ws.Output(0).type(), DALI_FLOAT); + ASSERT_EQ(ws.Output(1).type(), DALI_FLOAT); vector result1_cpu(tensor_elements); for (int sample_id = 0; sample_id < batch_size; sample_id++) { const auto *data0 = batch[0].tensor(sample_id); const auto *data1 = batch[1].tensor(sample_id); - auto *result0 = ws.OutputRef(0).tensor(sample_id); - auto *result1 = ws.OutputRef(1).tensor(sample_id); + auto *result0 = ws.Output(0).tensor(sample_id); + auto *result1 = ws.Output(1).tensor(sample_id); MemCopy(result1_cpu.data(), result1, tensor_elements * sizeof(float)); CUDA_CALL(cudaStreamSynchronize(0)); @@ -458,8 +458,8 @@ TEST(ArithmeticOpsTest, ConstantsPipeline) { for (int sample_id = 0; sample_id < batch_size; sample_id++) { const auto *data = batch.tensor(sample_id); - auto *result0 = ws.OutputRef(0).tensor(sample_id); - auto *result1 = ws.OutputRef(1).tensor(sample_id); + auto *result0 = ws.Output(0).tensor(sample_id); + auto *result1 = ws.Output(1).tensor(sample_id); for (int i = 0; i < tensor_elements; i++) { EXPECT_EQ(result0[i], data[i] + magic_int); @@ -521,16 +521,16 @@ class ArithmeticOpsScalarTest : public ::testing::TestWithParam pipe.Outputs(&ws); - ASSERT_EQ(ws.OutputRef(0).shape(), result_shape); - ASSERT_EQ(ws.OutputRef(1).shape(), result_shape); + ASSERT_EQ(ws.Output(0).shape(), result_shape); + ASSERT_EQ(ws.Output(1).shape(), result_shape); for (int sample_id = 0; sample_id < batch_size; sample_id++) { const auto *data0 = batch[0].tensor(sample_id); const auto *data1 = batch[1].tensor(sample_id); - const auto *result0 = ws.OutputRef(0).tensor(sample_id); - const auto *result1 = ws.OutputRef(1).tensor(sample_id); + const auto *result0 = ws.Output(0).tensor(sample_id); + const auto *result1 = ws.Output(1).tensor(sample_id); vector result1_cpu(result_shape[sample_id].num_elements()); @@ -666,8 +666,8 @@ TEST(ArithmeticOpsTest, UnaryPipeline) { vector result1_cpu(tensor_elements); for (int sample_id = 0; sample_id < batch_size; sample_id++) { - auto *result0 = ws.OutputRef(0).tensor(sample_id); - auto *result1 = ws.OutputRef(1).tensor(sample_id); + auto *result0 = ws.Output(0).tensor(sample_id); + auto *result1 = ws.Output(1).tensor(sample_id); MemCopy(result1_cpu.data(), result1, tensor_elements * sizeof(int)); CUDA_CALL(cudaStreamSynchronize(0)); diff --git a/dali/operators/math/expressions/expression_impl_factory.h b/dali/operators/math/expressions/expression_impl_factory.h index bf47ed34f1f..25a920df659 100644 --- a/dali/operators/math/expressions/expression_impl_factory.h +++ b/dali/operators/math/expressions/expression_impl_factory.h @@ -1,4 +1,4 @@ -// Copyright (c) 2019, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -130,19 +130,19 @@ struct ExprImplTask { }; inline OutputSamplePtr GetOutputSamplePointer(HostWorkspace &ws, int output_idx, int sample_idx) { - return ws.template OutputRef(output_idx)[sample_idx].raw_mutable_data(); + return ws.template Output(output_idx)[sample_idx].raw_mutable_data(); } inline OutputSamplePtr GetOutputSamplePointer(DeviceWorkspace &ws, int output_idx, int sample_idx) { - return ws.template OutputRef(output_idx).raw_mutable_tensor(sample_idx); + return ws.template Output(output_idx).raw_mutable_tensor(sample_idx); } inline InputSamplePtr GetInputSamplePointer(HostWorkspace &ws, int input_idx, int sample_idx) { - return ws.template InputRef(input_idx)[sample_idx].raw_data(); + return ws.template Input(input_idx)[sample_idx].raw_data(); } inline InputSamplePtr GetInputSamplePointer(DeviceWorkspace &ws, int input_idx, int sample_idx) { - return ws.template InputRef(input_idx).raw_tensor(sample_idx); + return ws.template Input(input_idx).raw_tensor(sample_idx); } template diff --git a/dali/operators/math/normalize/normalize.cc b/dali/operators/math/normalize/normalize.cc index bd8c02bb616..d779193e25a 100644 --- a/dali/operators/math/normalize/normalize.cc +++ b/dali/operators/math/normalize/normalize.cc @@ -146,7 +146,7 @@ using namespace normalize; // NOLINT template void Normalize::SetupTyped(const HostWorkspace &ws) { - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); int nsamples = input.num_samples(); int nthreads = ws.GetThreadPool().NumThreads(); @@ -248,11 +248,11 @@ template void Normalize::RunTyped(HostWorkspace &ws) { ThreadPool &tp = ws.GetThreadPool(); - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); TensorListView in_view = view(input); auto in_shape = input.shape(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); TensorListView out_view = view(output); output.SetLayout(input.GetLayout()); diff --git a/dali/operators/math/normalize/normalize.h b/dali/operators/math/normalize/normalize.h index 38ad4edd789..2e10259cab1 100644 --- a/dali/operators/math/normalize/normalize.h +++ b/dali/operators/math/normalize/normalize.h @@ -78,7 +78,7 @@ class NormalizeBase : public Operator { bool CanInferOutputs() const override { return true; } bool SetupImpl(std::vector &output_descs, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); data_shape_ = input.shape(); output_descs.resize(1); output_descs[0] = { data_shape_, output_type_ }; @@ -134,7 +134,7 @@ class NormalizeBase : public Operator { GetParamShapeFromAxes(); } else if (has_axis_names_arg_) { TensorLayout names = spec.GetArgument("axis_names"); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); auto dim_idx = GetDimIndices(input.GetLayout(), names); axes_ = dim_idx.to_vector(); SetAxisMask(); diff --git a/dali/operators/math/normalize/normalize_gpu.cu b/dali/operators/math/normalize/normalize_gpu.cu index 9897df915ea..035b819ca84 100644 --- a/dali/operators/math/normalize/normalize_gpu.cu +++ b/dali/operators/math/normalize/normalize_gpu.cu @@ -144,7 +144,7 @@ Normalize::BroadcastMean(KernelContext &ctx, float value) const { template void Normalize::SetupTyped(const DeviceWorkspace &ws) { - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); int nsamples = input.num_samples(); KernelContext ctx; @@ -207,10 +207,10 @@ void Normalize::SetupTyped(const DeviceWorkspace &ws) { template void Normalize::RunTyped(DeviceWorkspace &ws) { - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); TensorListView in_view = view(input); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); TensorListView out_view = view(output); output.SetLayout(input.GetLayout()); diff --git a/dali/operators/numba_function/numba_func.cc b/dali/operators/numba_function/numba_func.cc index 5bc92a90cea..bb359255900 100644 --- a/dali/operators/numba_function/numba_func.cc +++ b/dali/operators/numba_function/numba_func.cc @@ -189,7 +189,7 @@ bool NumbaFuncImpl::SetupImpl(std::vector &output_desc, output_desc.resize(out_types_.size()); in_shapes_.resize(ninputs); for (int in_id = 0; in_id < ninputs; in_id++) { - auto& in = ws.InputRef(in_id); + auto& in = ws.Input(in_id); in_shapes_[in_id] = in.shape(); DALI_ENFORCE(in_shapes_[in_id].sample_dim() == ins_ndim_[in_id], make_string( "Number of dimensions passed in `ins_ndim` at index ", in_id, @@ -210,7 +210,7 @@ bool NumbaFuncImpl::SetupImpl(std::vector &output_desc, if (!setup_fn_) { for (int i = 0; i < noutputs; i++) { - const auto &in = ws.InputRef(i); + const auto &in = ws.Input(i); output_desc[i] = {in.shape(), in.type()}; } return true; @@ -251,20 +251,20 @@ bool NumbaFuncImpl::SetupImpl(std::vector &output_desc, template <> void NumbaFuncImpl::RunImpl(workspace_t &ws) { - auto N = ws.InputRef(0).shape().num_samples(); + auto N = ws.Input(0).shape().num_samples(); std::vector out_ptrs; std::vector in_ptrs; out_ptrs.resize(N * out_types_.size()); in_ptrs.resize(N * in_types_.size()); for (size_t out_id = 0; out_id < out_types_.size(); out_id++) { - auto& out = ws.OutputRef(out_id); + auto& out = ws.Output(out_id); for (int i = 0; i < N; i++) { out_ptrs[N * out_id + i] = reinterpret_cast(out[i].raw_mutable_data()); } } for (size_t in_id = 0; in_id < in_types_.size(); in_id++) { - auto& in = ws.InputRef(in_id); + auto& in = ws.Input(in_id); for (int i = 0; i < N; i++) { in_ptrs[N * in_id + i] = reinterpret_cast(in[i].raw_mutable_data()); } @@ -281,7 +281,7 @@ void NumbaFuncImpl::RunImpl(workspace_t &ws) { return; } - auto &out = ws.OutputRef(0); + auto &out = ws.Output(0); auto out_shape = out.shape(); auto &tp = ws.GetThreadPool(); for (int sample_id = 0; sample_id < N; sample_id++) { diff --git a/dali/operators/python_function/dltensor_function.cc b/dali/operators/python_function/dltensor_function.cc index 845ae62185c..3674357d60c 100644 --- a/dali/operators/python_function/dltensor_function.cc +++ b/dali/operators/python_function/dltensor_function.cc @@ -78,7 +78,7 @@ py::list PrepareDLTensorInputs(HostWorkspace &ws) { for (Index idx = 0; idx < ws.NumInput(); ++idx) { py::list dl_tensor_list; for (Index i = 0; i < ws.GetInputBatchSize(idx); ++i) { - auto &t = ws.InputRef(idx)[i]; + auto &t = ws.Input(idx)[i]; auto dl_capsule = TensorToDLPackView(const_cast&>(t)); dl_tensor_list.append(dl_capsule); } @@ -91,7 +91,7 @@ template <> py::list PrepareDLTensorInputs(DeviceWorkspace &ws) { py::list input_tuple; for (Index idx = 0; idx < ws.NumInput(); ++idx) { - auto &tlist = ws.InputRef(idx); + auto &tlist = ws.Input(idx); py::list dl_tensor_list = TensorListToDLPackView(tlist); input_tuple.append(dl_tensor_list); } @@ -106,7 +106,7 @@ py::list PrepareDLTensorInputsPerSample(HostWorkspace &ws) { for (Index s = 0; s < batch_size; ++s) { py::list tuple; for (Index idx = 0; idx < ws.NumInput(); ++idx) { - auto &t = ws.InputRef(idx)[s]; + auto &t = ws.Input(idx)[s]; auto dl_capsule = TensorToDLPackView(const_cast&>(t)); tuple.append(dl_capsule); } @@ -119,10 +119,10 @@ template <> py::list PrepareDLTensorInputsPerSample(DeviceWorkspace &ws) { std::vector input_tuples; if (ws.NumInput() == 0) return py::cast(input_tuples); - Index batch_size = ws.InputRef(0).num_samples(); + Index batch_size = ws.Input(0).num_samples(); input_tuples.resize(batch_size); for (Index idx = 0; idx < ws.NumInput(); ++idx) { - py::list dl_tensor_list = TensorListToDLPackView(ws.InputRef(idx)); + py::list dl_tensor_list = TensorListToDLPackView(ws.Input(idx)); for (Index s = 0; s < batch_size; ++s) { input_tuples[s].append(dl_tensor_list[s]); } diff --git a/dali/operators/python_function/dltensor_function.h b/dali/operators/python_function/dltensor_function.h index c8d2c6ddf8c..777574f7479 100644 --- a/dali/operators/python_function/dltensor_function.h +++ b/dali/operators/python_function/dltensor_function.h @@ -108,7 +108,7 @@ void PrepareOutputs(workspace_t &ws, const py::object &output_o, int ba py::list dl_list = py::cast(return_tuple[idx]); auto dl_tensors = CastToDLTensorList(dl_list, batch_size, idx); if (dl_tensors.empty()) continue; - auto &tlist = ws.template OutputRef(idx); + auto &tlist = ws.template Output(idx); tlist.Resize(GetDLTensorListShape(dl_tensors), DLToDALIType(dl_tensors[0]->dl_tensor.dtype)); CopyOutputData(tlist, dl_tensors, batch_size, ws); } @@ -232,7 +232,7 @@ class DLTensorPythonFunctionImpl : public Operator { void SetOutputLayouts(workspace_t &ws) { Index output_idx = 0; for (auto layout : output_layouts_) { - auto &output = ws.template OutputRef(output_idx); + auto &output = ws.template Output(output_idx); output.SetLayout(layout); ++output_idx; } diff --git a/dali/operators/random/batch_permutation.cc b/dali/operators/random/batch_permutation.cc index 98a03d433fd..37de42f3f61 100644 --- a/dali/operators/random/batch_permutation.cc +++ b/dali/operators/random/batch_permutation.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -31,7 +31,7 @@ indexing samples in the batch.)") points, that is ``out[i] != i``. This argument is ignored when batch size is 1.)", false); void BatchPermutation::RunImpl(HostWorkspace &ws) { - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); int N = ws.GetRequestedBatchSize(0); if (N < 1) return; diff --git a/dali/operators/random/rng_base.h b/dali/operators/random/rng_base.h index 20ba22e40d3..bb00a9af2bd 100644 --- a/dali/operators/random/rng_base.h +++ b/dali/operators/random/rng_base.h @@ -46,7 +46,7 @@ class RNGBase : public Operator { int GetBatchSize(const workspace_t &ws) const { if (spec_.NumRegularInput() == 1) - return ws.template InputRef(0).shape().size(); + return ws.template Input(0).shape().size(); else return ws.GetRequestedBatchSize(0); } @@ -54,7 +54,7 @@ class RNGBase : public Operator { bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { if (IsNoiseGen) - dtype_ = ws.template InputRef(0).type(); + dtype_ = ws.template Input(0).type(); else if (!spec_.TryGetArgument(dtype_, "dtype")) dtype_ = This().DefaultDataType(); @@ -65,13 +65,13 @@ class RNGBase : public Operator { "Providing argument \"shape\" is incompatible with providing a shape-like input"); if (IsNoiseGen) { - shape_ = ws.template InputRef(0).shape(); + shape_ = ws.template Input(0).shape(); } else if (has_shape_like) { if (ws.template InputIsType(0)) { - shape_ = ws.template InputRef(0).shape(); + shape_ = ws.template Input(0).shape(); } else if (std::is_same::value && ws.template InputIsType(0)) { - shape_ = ws.template InputRef(0).shape(); + shape_ = ws.template Input(0).shape(); } else { DALI_FAIL( "Shape-like input can be either CPUBackend or GPUBackend for case of GPU operators."); diff --git a/dali/operators/random/rng_base_cpu.h b/dali/operators/random/rng_base_cpu.h index b2e164fe70e..15743f1bfcc 100644 --- a/dali/operators/random/rng_base_cpu.h +++ b/dali/operators/random/rng_base_cpu.h @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -109,7 +109,7 @@ template void RNGBase::RunImplTyped(workspace_t &ws) { // Should never be called for Backend != CPUBackend static_assert(std::is_same::value, "Invalid backend"); - auto &output = ws.template OutputRef(0); + auto &output = ws.template Output(0); auto out_view = view(output); const auto &out_shape = out_view.shape; auto &tp = ws.GetThreadPool(); @@ -121,7 +121,7 @@ void RNGBase::RunImplTyped(workspace_t &w TensorListView, const T, DynamicDimensions> in_view; if (IsNoiseGen) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); in_view = view(input); output.SetLayout(input.GetLayout()); } diff --git a/dali/operators/random/rng_base_gpu.cuh b/dali/operators/random/rng_base_gpu.cuh index e579c84e951..fee0bc986ce 100644 --- a/dali/operators/random/rng_base_gpu.cuh +++ b/dali/operators/random/rng_base_gpu.cuh @@ -126,7 +126,7 @@ template template void RNGBase::RunImplTyped(workspace_t &ws) { static_assert(std::is_same::value, "Unexpected backend"); - auto &output = ws.template OutputRef(0); + auto &output = ws.template Output(0); auto rngs = backend_data_.randomizer_.states(); int block_sz = backend_data_.block_size_; int max_nblocks = backend_data_.max_blocks_; @@ -139,7 +139,7 @@ void RNGBase::RunImplTyped(workspace_t &w } if (IsNoiseGen) { - const auto& input = ws.template InputRef(0); + const auto& input = ws.template Input(0); in_view = view(input); output.SetLayout(input.GetLayout()); } diff --git a/dali/operators/reader/coco_reader_op.cc b/dali/operators/reader/coco_reader_op.cc index 62d46589acf..aaafb98702c 100755 --- a/dali/operators/reader/coco_reader_op.cc +++ b/dali/operators/reader/coco_reader_op.cc @@ -201,7 +201,7 @@ void COCOReader::RunImpl(SampleWorkspace &ws) { const ImageLabelWrapper& image_label = GetSample(ws.data_idx()); Index image_size = image_label.image.size(); - auto &image_output = ws.OutputRef(0); + auto &image_output = ws.Output(0); int image_idx = image_label.label; image_output.Resize({image_size}, DALI_UINT8); @@ -210,20 +210,20 @@ void COCOReader::RunImpl(SampleWorkspace &ws) { auto &loader_impl = LoaderImpl(); auto bboxes = loader_impl.bboxes(image_idx); - auto &boxes_output = ws.OutputRef(1); + auto &boxes_output = ws.Output(1); boxes_output.Resize({bboxes.size(), 4}, DALI_FLOAT); std::memcpy(boxes_output.mutable_data(), bboxes.data(), bboxes.size() * sizeof(vec<4>)); auto labels = loader_impl.labels(image_idx); - auto &labels_output = ws.OutputRef(2); + auto &labels_output = ws.Output(2); labels_output.Resize({labels.size()}, DALI_INT32); // 0.28dev: changed shape from {N, 1} to {N} std::memcpy(labels_output.mutable_data(), labels.data(), labels.size() * sizeof(int)); int curr_out_idx = 3; if (output_polygon_masks_) { - auto &polygons_output = ws.OutputRef(curr_out_idx++); + auto &polygons_output = ws.Output(curr_out_idx++); auto polygons = loader_impl.polygons(image_idx); polygons_output.Resize({polygons.size(), 3}, DALI_INT32); std::memcpy(polygons_output.mutable_data(), @@ -235,7 +235,7 @@ void COCOReader::RunImpl(SampleWorkspace &ws) { poly_data[i * 3 + 2] *= 2; } } - auto &vertices_output = ws.OutputRef(curr_out_idx++); + auto &vertices_output = ws.Output(curr_out_idx++); auto vertices = loader_impl.vertices(image_idx); vertices_output.Resize({vertices.size(), 2}, DALI_FLOAT); std::memcpy(vertices_output.mutable_data(), @@ -243,7 +243,7 @@ void COCOReader::RunImpl(SampleWorkspace &ws) { } if (output_pixelwise_masks_) { - auto &masks_output = ws.OutputRef(curr_out_idx++); + auto &masks_output = ws.Output(curr_out_idx++); auto masks_info = loader_impl.pixelwise_masks_info(image_idx); masks_output.Resize(masks_info.shape, DALI_INT32); masks_output.SetLayout("HWC"); @@ -251,7 +251,7 @@ void COCOReader::RunImpl(SampleWorkspace &ws) { } if (output_image_ids_) { - auto &id_output = ws.OutputRef(curr_out_idx++); + auto &id_output = ws.Output(curr_out_idx++); id_output.Resize({1}, DALI_INT32); *(id_output.mutable_data()) = loader_impl.image_id(image_idx); } diff --git a/dali/operators/reader/coco_reader_op_test.cc b/dali/operators/reader/coco_reader_op_test.cc index 84dda50f870..6b86bfe0051 100644 --- a/dali/operators/reader/coco_reader_op_test.cc +++ b/dali/operators/reader/coco_reader_op_test.cc @@ -107,7 +107,7 @@ class CocoReaderTest : public ::testing::Test { int ObjectCount(bool masks) { return masks ? 7 : 194; } std::vector CopyIds(DeviceWorkspace &ws, int ids_out_idx = 3) { - auto &output = ws.OutputRef(ids_out_idx); + auto &output = ws.Output(ids_out_idx); const auto &shape = output.shape(); vector ids(shape.size()); @@ -182,8 +182,8 @@ class CocoReaderTest : public ::testing::Test { void CheckInstances(DeviceWorkspace &ws, bool ltrb, bool ratio, bool skip_empty, int expected_size, bool polygon_masks, bool polygon_masks_legacy) { - const auto &boxes_output = ws.OutputRef(1); - const auto &labels_output = ws.OutputRef(2); + const auto &boxes_output = ws.Output(1); + const auto &labels_output = ws.Output(2); const auto &boxes_shape = boxes_output.shape(); const auto &labels_shape = labels_output.shape(); @@ -254,8 +254,8 @@ class CocoReaderTest : public ::testing::Test { } if (polygon_masks || polygon_masks_legacy) { - const auto &polygons_output = ws.OutputRef(3); - const auto &vertices_output = ws.OutputRef(4); + const auto &polygons_output = ws.Output(3); + const auto &vertices_output = ws.Output(4); const auto &polygons_shape = polygons_output.shape(); const auto &vertices_shape = vertices_output.shape(); @@ -605,7 +605,7 @@ TEST_F(CocoReaderTest, PixelwiseMasks) { pipe2.Outputs(&ws2); for (auto *ws : {&ws1, &ws2}) { - const auto &masks_output = ws->OutputRef(3); + const auto &masks_output = ws->Output(3); const auto &masks_shape = masks_output.shape(); TensorListShape<3> pixelwise_masks_shape({ diff --git a/dali/operators/reader/file_reader_op.h b/dali/operators/reader/file_reader_op.h index 8cd41e72568..23b9f0d4d08 100644 --- a/dali/operators/reader/file_reader_op.h +++ b/dali/operators/reader/file_reader_op.h @@ -37,8 +37,8 @@ class FileReader : public DataReader { const auto& image_label = GetSample(idx); // copy from raw_data -> outputs directly - auto &image_output = ws.OutputRef(0); - auto &label_output = ws.OutputRef(1); + auto &image_output = ws.Output(0); + auto &label_output = ws.Output(1); Index image_size = image_label.image.size(); diff --git a/dali/operators/reader/nemo_asr_reader_op.cc b/dali/operators/reader/nemo_asr_reader_op.cc index 0d2bc7472ff..02ad5280411 100755 --- a/dali/operators/reader/nemo_asr_reader_op.cc +++ b/dali/operators/reader/nemo_asr_reader_op.cc @@ -193,7 +193,7 @@ void NemoAsrReader::RunImpl(SampleWorkspace &ws) { const auto &sample = GetSample(ws.data_idx()); const auto &sample_audio = GetDecodedAudioSample(ws.data_idx()); - auto &audio = ws.OutputRef(0); + auto &audio = ws.Output(0); audio.Copy(sample_audio, 0); DALIMeta meta; @@ -203,14 +203,14 @@ void NemoAsrReader::RunImpl(SampleWorkspace &ws) { int next_out_idx = 1; if (read_sr_) { - auto &sample_rate = ws.OutputRef(next_out_idx++); + auto &sample_rate = ws.Output(next_out_idx++); sample_rate.Resize({}, DALI_FLOAT); sample_rate.mutable_data()[0] = sample.audio_meta().sample_rate; sample_rate.SetMeta(meta); } if (read_text_) { - auto &text_out = ws.OutputRef(next_out_idx++); + auto &text_out = ws.Output(next_out_idx++); const auto &text = sample.text(); int64_t text_sz = text.length(); text_out.Resize({text_sz}, DALI_UINT8); @@ -219,7 +219,7 @@ void NemoAsrReader::RunImpl(SampleWorkspace &ws) { } if (read_idxs_) { - auto &idxs_out = ws.OutputRef(next_out_idx++); + auto &idxs_out = ws.Output(next_out_idx++); idxs_out.Resize({1}, DALI_INT64); *idxs_out.mutable_data() = sample.index(); idxs_out.SetMeta(meta); diff --git a/dali/operators/reader/numpy_reader_gpu_op_impl.cu b/dali/operators/reader/numpy_reader_gpu_op_impl.cu index 0c557a4cbe9..0a7e333d6c0 100644 --- a/dali/operators/reader/numpy_reader_gpu_op_impl.cu +++ b/dali/operators/reader/numpy_reader_gpu_op_impl.cu @@ -26,7 +26,7 @@ namespace dali { template void NumpyReaderGPU::RunImplTyped(DeviceWorkspace &ws) { - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); const auto &out_sh = output.shape(); int ndim = out_sh.sample_dim(); int nsamples = out_sh.num_samples(); @@ -152,7 +152,7 @@ void NumpyReaderGPU::RunImplTyped(DeviceWorkspace &ws) { } void NumpyReaderGPU::RunImpl(DeviceWorkspace &ws) { - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); int ndim = output.shape().sample_dim(); auto dtype = output.type(); VALUE_SWITCH(ndim, Dims, NUMPY_ALLOWED_DIMS, ( diff --git a/dali/operators/reader/numpy_reader_op.cc b/dali/operators/reader/numpy_reader_op.cc index 4fe68206ef9..fe91ed2fdc1 100644 --- a/dali/operators/reader/numpy_reader_op.cc +++ b/dali/operators/reader/numpy_reader_op.cc @@ -230,7 +230,7 @@ submodule and renamed to follow a common pattern. This is a placeholder operator functionality to allow for backward compatibility.)code"); // Deprecated in 1.0; void NumpyReaderCPU::RunImpl(HostWorkspace &ws) { - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); const auto &out_sh = output.shape(); int nsamples = out_sh.num_samples(); auto &thread_pool = ws.GetThreadPool(); diff --git a/dali/operators/reader/parser/caffe2_parser.h b/dali/operators/reader/parser/caffe2_parser.h index fa2317e3848..1edfb29e1f1 100644 --- a/dali/operators/reader/parser/caffe2_parser.h +++ b/dali/operators/reader/parser/caffe2_parser.h @@ -121,19 +121,19 @@ void ParseLabels(const caffe2::TensorProtos& protos, const int num_labels, SampleWorkspace* ws, int consumed_inputs) { - auto& label_tensor = ws->OutputRef(consumed_inputs); + auto& label_tensor = ws->Output(consumed_inputs); switch (label_type) { case SINGLE_LABEL: { // single element, from protos(1) to Output(consumed_inputs) // ensure we only have a single label in the proto DALI_ENFORCE(proto_data_size(protos.protos(consumed_inputs)) == 1); - extract_data(protos.protos(consumed_inputs), ws->OutputRef(consumed_inputs)); + extract_data(protos.protos(consumed_inputs), ws->Output(consumed_inputs)); break; } case MULTI_LABEL_SPARSE: { // multiple labels, all 1. in elements defined in protos(consumed_inputs) - auto& label_tensor = ws->OutputRef(consumed_inputs); + auto& label_tensor = ws->Output(consumed_inputs); label_tensor.set_type(); label_tensor.Resize({num_labels}); @@ -150,7 +150,7 @@ void ParseLabels(const caffe2::TensorProtos& protos, } case MULTI_LABEL_DENSE: { // multiple elements, stored contiguously - extract_data(protos.protos(consumed_inputs), ws->OutputRef(consumed_inputs)); + extract_data(protos.protos(consumed_inputs), ws->Output(consumed_inputs)); break; } case MULTI_LABEL_WEIGHTED_SPARSE: { @@ -196,7 +196,7 @@ class Caffe2Parser : public Parser> { if (image_available_) { - auto& image = ws->OutputRef(consumed_inputs); + auto& image = ws->Output(consumed_inputs); const caffe2::TensorProto& image_proto = protos.protos(consumed_inputs); // copy image -- if type is string, image is encoded, if bytes, image isn't encoded @@ -243,7 +243,7 @@ class Caffe2Parser : public Parser> { for (int i = additional_proto_start; i < additional_proto_end; ++i) { auto& additional_proto = protos.protos(i); - auto& output_tensor = ws->OutputRef(consumed_inputs); + auto& output_tensor = ws->Output(consumed_inputs); switch (additional_proto.data_type()) { case caffe2::TensorProto::FLOAT: @@ -269,7 +269,7 @@ class Caffe2Parser : public Parser> { DALI_ENFORCE(bbox_proto.data_type() == caffe2::TensorProto::INT32); DALI_ENFORCE(bbox_proto.int32_data_size() == 4); - extract_data(bbox_proto, ws->OutputRef(consumed_inputs)); + extract_data(bbox_proto, ws->Output(consumed_inputs)); } } diff --git a/dali/operators/reader/parser/caffe_parser.h b/dali/operators/reader/parser/caffe_parser.h index 6a68cda7ac1..bca229cea07 100644 --- a/dali/operators/reader/parser/caffe_parser.h +++ b/dali/operators/reader/parser/caffe_parser.h @@ -36,7 +36,7 @@ class CaffeParser : public Parser> { if (image_available_ && datum.has_data()) { bool encoded_data = true; - auto& image = ws->OutputRef(out_tensors); + auto& image = ws->Output(out_tensors); if (datum.has_encoded() && !datum.encoded()) { encoded_data = false; } @@ -53,7 +53,7 @@ class CaffeParser : public Parser> { } if (label_available_ && datum.has_label()) { - auto& label = ws->OutputRef(out_tensors); + auto& label = ws->Output(out_tensors); // copy label label.Resize({1}, DALI_INT32); diff --git a/dali/operators/reader/parser/parser_test.cc b/dali/operators/reader/parser/parser_test.cc index fb4723ab1d7..bc788424ca4 100644 --- a/dali/operators/reader/parser/parser_test.cc +++ b/dali/operators/reader/parser/parser_test.cc @@ -45,7 +45,7 @@ class IntArrayParser : public Parser { printf("H: %d, W: %d, C: %d\n", H, W, C); - Tensor& output = ws->template OutputRef(0); + Tensor& output = ws->template Output(0); output.Resize({H, W, C}, DALI_INT32); int *output_data = output.template mutable_data(); diff --git a/dali/operators/reader/parser/recordio_parser.h b/dali/operators/reader/parser/recordio_parser.h index f6d9daeb329..02258ef4b73 100644 --- a/dali/operators/reader/parser/recordio_parser.h +++ b/dali/operators/reader/parser/recordio_parser.h @@ -35,8 +35,8 @@ class RecordIOParser : public Parser> { } void Parse(const Tensor& data, SampleWorkspace* ws) override { - auto& image = ws->OutputRef(0); - auto& label = ws->OutputRef(1); + auto& image = ws->Output(0); + auto& label = ws->Output(1); ReadSingleImageRecordIO(image, label, data.data()); image.SetSourceInfo(data.GetSourceInfo()); } diff --git a/dali/operators/reader/parser/sequence_parser.cc b/dali/operators/reader/parser/sequence_parser.cc index 0f02c2f6f3c..0f76cdaeef4 100644 --- a/dali/operators/reader/parser/sequence_parser.cc +++ b/dali/operators/reader/parser/sequence_parser.cc @@ -21,7 +21,7 @@ namespace dali { void SequenceParser::Parse(const TensorSequence& data, SampleWorkspace* ws) { - auto& sequence = ws->OutputRef(0); + auto& sequence = ws->Output(0); sequence.SetLayout("FHWC"); Index seq_length = data.tensors.size(); diff --git a/dali/operators/reader/parser/tfrecord_parser.h b/dali/operators/reader/parser/tfrecord_parser.h index c38450f5ca6..a9d6c5444df 100644 --- a/dali/operators/reader/parser/tfrecord_parser.h +++ b/dali/operators/reader/parser/tfrecord_parser.h @@ -63,7 +63,7 @@ class TFRecordParser : public Parser> { " (raw data length: ", length, "bytes).")); for (size_t i = 0; i < features_.size(); ++i) { - auto& output = ws->OutputRef(i); + auto& output = ws->Output(i); Feature& f = features_[i]; std::string& name = feature_names_[i]; auto& feature = example.features().feature(); diff --git a/dali/operators/reader/reader_op.h b/dali/operators/reader/reader_op.h index accd945b681..bc5c9c65f5f 100644 --- a/dali/operators/reader/reader_op.h +++ b/dali/operators/reader/reader_op.h @@ -152,7 +152,7 @@ class DataReader : public Operator { void EnforceUniformOutput(const HostWorkspace &ws) const { for (int out_idx = 0; out_idx < ws.NumOutput(); out_idx++) { - auto &out = ws.OutputRef(out_idx); + auto &out = ws.Output(out_idx); int n = out.num_samples(); if (n < 2) continue; @@ -237,7 +237,7 @@ class DataReader : public Operator { DALI_ENFORCE(cached_outputs.size() == num_outputs, "Unexpected number of outputs"); for (std::size_t i = 0; i < cached_outputs.size(); i++) { - auto& output = ws->OutputRef(i); + auto& output = ws->Output(i); output.Copy(cached_outputs[i], 0); } return; @@ -262,7 +262,7 @@ class DataReader : public Operator { first_output.Resize({1}, DALI_UINT8); for (std::size_t i = 1; i < cached_outputs.size(); i++) { - auto& output = ws->OutputRef(i); + auto& output = ws->Output(i); cached_outputs[i].set_pinned(false); cached_outputs[i].Copy(output, 0); } diff --git a/dali/operators/reader/reader_op_test.cc b/dali/operators/reader/reader_op_test.cc index 231a1d0ba44..fd25495d242 100644 --- a/dali/operators/reader/reader_op_test.cc +++ b/dali/operators/reader/reader_op_test.cc @@ -88,7 +88,7 @@ class DummyDataReader : public DataReader> { void RunImpl(SampleWorkspace &ws) override { std::this_thread::sleep_for(std::chrono::milliseconds(5)); - ws.OutputRef(0).Copy(GetSample(ws.data_idx()), 0); + ws.Output(0).Copy(GetSample(ws.data_idx()), 0); } private: @@ -220,7 +220,7 @@ TYPED_TEST(ReaderTest, SequenceTest) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - auto shape = ws.OutputRef(0).AsTensor()->shape(); + auto shape = ws.Output(0).AsTensor()->shape(); // We have NFHWC format const auto batch_size = shape[0]; const auto frame_count = shape[1]; @@ -236,7 +236,7 @@ TYPED_TEST(ReaderTest, SequenceTest) { auto start_frame = (i * batch_size + sample) % (16 - 3 + 1); for (int frame = 0; frame < frame_count; frame++) { auto off = sample * seq_size + frame * frame_size; - auto val = ws.OutputRef(0).AsTensor()->data()[off]; + auto val = ws.Output(0).AsTensor()->data()[off]; decltype(val) expected = start_frame + frame; ASSERT_EQ(val, expected); } diff --git a/dali/operators/reader/video_reader_decoder_op.cc b/dali/operators/reader/video_reader_decoder_op.cc index 1548a08c04f..664e693937a 100644 --- a/dali/operators/reader/video_reader_decoder_op.cc +++ b/dali/operators/reader/video_reader_decoder_op.cc @@ -26,12 +26,12 @@ VideoReaderDecoder::VideoReaderDecoder(const OpSpec &spec) void VideoReaderDecoder::RunImpl(SampleWorkspace &ws) { const auto &sample = GetSample(ws.data_idx()); - auto &video_output = ws.OutputRef(0); + auto &video_output = ws.Output(0); video_output.Copy(sample.data_, 0); if (has_labels_) { - auto &label_output = ws.OutputRef(1); + auto &label_output = ws.Output(1); label_output.Resize({}, DALIDataType::DALI_INT32); label_output.mutable_data()[0] = sample.label_; } diff --git a/dali/operators/reader/video_reader_decoder_op_test.cc b/dali/operators/reader/video_reader_decoder_op_test.cc index 47370a765fe..2394430677d 100644 --- a/dali/operators/reader/video_reader_decoder_op_test.cc +++ b/dali/operators/reader/video_reader_decoder_op_test.cc @@ -65,8 +65,8 @@ TEST_F(VideoReaderDecoderTest, CpuConstantFrameRate) { pipe.RunGPU(); pipe.Outputs(&ws); - auto &frame_video_output = ws.template OutputRef(0); - auto &frame_label_output = ws.template OutputRef(1); + auto &frame_video_output = ws.template Output(0); + auto &frame_label_output = ws.template Output(1); for (int sample_id = 0; sample_id < batch_size; ++sample_id) { const auto sample = frame_video_output.tensor(sample_id); @@ -133,8 +133,8 @@ TEST_F(VideoReaderDecoderTest, CpuVariableFrameRate) { pipe.RunGPU(); pipe.Outputs(&ws); - auto &frame_video_output = ws.template OutputRef(0); - auto &frame_label_output = ws.template OutputRef(1); + auto &frame_video_output = ws.template Output(0); + auto &frame_label_output = ws.template Output(1); for (int sample_id = 0; sample_id < batch_size; ++sample_id) { const auto sample = frame_video_output.tensor(sample_id); @@ -194,7 +194,7 @@ TEST_F(VideoReaderDecoderTest, RandomShuffle) { pipe.RunGPU(); pipe.Outputs(&ws); - auto &frame_video_output = ws.template OutputRef(0); + auto &frame_video_output = ws.template Output(0); const auto sample = frame_video_output.tensor(0); CompareFrames(sample, this->GetCfrFrame(0, expected_order[sequence_id]), this->FrameSize(0)); } @@ -268,11 +268,11 @@ TEST_F(VideoReaderDecoderTest, CompareReaders) { pipe.RunGPU(); pipe.Outputs(&ws); - auto &frame_video_output = ws.template OutputRef(0); - auto &frame_gpu_video_output = ws.template OutputRef(1); + auto &frame_video_output = ws.template Output(0); + auto &frame_gpu_video_output = ws.template Output(1); - auto &frame_label_output = ws.template OutputRef(2); - auto &frame_gpu_label_output = ws.template OutputRef(3); + auto &frame_label_output = ws.template Output(2); + auto &frame_gpu_label_output = ws.template Output(3); for (int sample_id = 0; sample_id < batch_size; ++sample_id) { const auto sample = frame_video_output.tensor(sample_id); diff --git a/dali/operators/reader/video_reader_op.h b/dali/operators/reader/video_reader_op.h index a00032912c7..5e7d9a73697 100644 --- a/dali/operators/reader/video_reader_op.h +++ b/dali/operators/reader/video_reader_op.h @@ -129,16 +129,16 @@ class VideoReader : public DataReader { void PrepareAdditionalOutputs(DeviceWorkspace &ws) { int output_index = 1; if (output_labels_) { - label_output_ = &ws.OutputRef(output_index++); + label_output_ = &ws.Output(output_index++); label_output_->Resize(label_shape_, DALI_INT32); if (can_use_frames_timestamps_) { if (enable_frame_num_) { - frame_num_output_ = &ws.OutputRef(output_index++); + frame_num_output_ = &ws.Output(output_index++); frame_num_output_->Resize(frame_num_shape_, DALI_INT32); } if (enable_timestamps_) { - timestamp_output_ = &ws.OutputRef(output_index++); + timestamp_output_ = &ws.Output(output_index++); timestamp_output_->Resize(timestamp_shape_, DALI_FLOAT64); } } @@ -181,7 +181,7 @@ class VideoReader : public DataReader { } void RunImpl(DeviceWorkspace &ws) override { - auto &video_output = ws.OutputRef(0); + auto &video_output = ws.Output(0); auto &curent_batch = prefetched_batch_tensors_[curr_batch_consumer_]; SetOutputShapeType(video_output, ws); diff --git a/dali/operators/reader/video_reader_op_test.cc b/dali/operators/reader/video_reader_op_test.cc index 4a79b1273fc..11b7d3f5692 100644 --- a/dali/operators/reader/video_reader_op_test.cc +++ b/dali/operators/reader/video_reader_op_test.cc @@ -87,7 +87,7 @@ TEST_F(VideoReaderTest, ConstantFrameRate) { pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), 1); @@ -139,8 +139,8 @@ TEST_F(VideoReaderTest, MultipleVideoResolution) { pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_output = ws.OutputRef(0); - const auto &labels_output = ws.OutputRef(1); + const auto &frames_output = ws.Output(0); + const auto &labels_output = ws.Output(1); TensorList labels_cpu; labels_cpu.Copy(labels_output, 0); @@ -196,7 +196,7 @@ TEST_F(VideoReaderTest, PackedBFrames) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), batch_size); @@ -243,7 +243,7 @@ TEST_F(VideoReaderTest, Vp9Profile0) { } } - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), 1); @@ -282,7 +282,7 @@ TEST_F(VideoReaderTest, Vp9Profile2) { } } - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), 1); @@ -319,7 +319,7 @@ TEST_F(VideoReaderTest, Vp8Profile0) { } } - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), 1); @@ -358,7 +358,7 @@ TEST_F(VideoReaderTest, MJpeg) { } } - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), 1); @@ -399,7 +399,7 @@ TEST_F(VideoReaderTest, HEVC) { } } - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), 16); @@ -430,9 +430,9 @@ TEST_F(VideoReaderTest, FrameLabels) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_gpu = ws.OutputRef(0); - const auto &label_gpu = ws.OutputRef(1); - const auto &frame_num_gpu = ws.OutputRef(2); + const auto &frames_gpu = ws.Output(0); + const auto &label_gpu = ws.Output(1); + const auto &frame_num_gpu = ws.Output(2); TensorList frames_cpu; frames_cpu.Copy(frames_gpu, 0); @@ -476,9 +476,9 @@ TEST_F(VideoReaderTest, FrameLabelsFilenames) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_gpu = ws.OutputRef(0); - const auto &label_gpu = ws.OutputRef(1); - const auto &frame_num_gpu = ws.OutputRef(2); + const auto &frames_gpu = ws.Output(0); + const auto &label_gpu = ws.Output(1); + const auto &frame_num_gpu = ws.Output(2); TensorList frames_cpu; frames_cpu.Copy(frames_gpu, 0); @@ -524,9 +524,9 @@ TEST_F(VideoReaderTest, LabelsFilenames) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_gpu = ws.OutputRef(0); - const auto &label_gpu = ws.OutputRef(1); - const auto &frame_num_gpu = ws.OutputRef(2); + const auto &frames_gpu = ws.Output(0); + const auto &label_gpu = ws.Output(1); + const auto &frame_num_gpu = ws.Output(2); TensorList frames_cpu; frames_cpu.Copy(frames_gpu, 0); @@ -572,10 +572,10 @@ TEST_F(VideoReaderTest, FrameLabelsWithFileListFrameNum) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_gpu = ws.OutputRef(0); - const auto &label_gpu = ws.OutputRef(1); - const auto &frame_num_gpu = ws.OutputRef(2); - const auto ×tamp_gpu = ws.OutputRef(3); + const auto &frames_gpu = ws.Output(0); + const auto &label_gpu = ws.Output(1); + const auto &frame_num_gpu = ws.Output(2); + const auto ×tamp_gpu = ws.Output(3); TensorList frames_cpu; frames_cpu.Copy(frames_gpu, 0); @@ -634,10 +634,10 @@ TEST_F(VideoReaderTest, TimestampLabels) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_gpu = ws.OutputRef(0); - const auto &label_gpu = ws.OutputRef(1); - const auto &frame_num_gpu = ws.OutputRef(2); - const auto ×tamp_gpu = ws.OutputRef(3); + const auto &frames_gpu = ws.Output(0); + const auto &label_gpu = ws.Output(1); + const auto &frame_num_gpu = ws.Output(2); + const auto ×tamp_gpu = ws.Output(3); TensorList frames_cpu; frames_cpu.Copy(frames_gpu, 0); @@ -680,9 +680,9 @@ TEST_F(VideoReaderTest, StartEndLabels) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_gpu = ws.OutputRef(0); - const auto &label_gpu = ws.OutputRef(1); - const auto &frame_num_gpu = ws.OutputRef(2); + const auto &frames_gpu = ws.Output(0); + const auto &label_gpu = ws.Output(1); + const auto &frame_num_gpu = ws.Output(2); TensorList frames_cpu; frames_cpu.Copy(frames_gpu, 0); @@ -720,7 +720,7 @@ TEST_F(VideoReaderTest, MultipleFrameRates) { pipe.RunCPU(); pipe.RunGPU(); pipe.Outputs(&ws); - const auto &frames_output = ws.OutputRef(0); + const auto &frames_output = ws.Output(0); const auto &frames_shape = frames_output.shape(); ASSERT_EQ(frames_shape.size(), batch_size); diff --git a/dali/operators/reader/webdataset_reader_op.cc b/dali/operators/reader/webdataset_reader_op.cc index 3a9d0237a0b..bf321c804ee 100644 --- a/dali/operators/reader/webdataset_reader_op.cc +++ b/dali/operators/reader/webdataset_reader_op.cc @@ -47,7 +47,7 @@ void WebdatasetReader::RunImpl(HostWorkspace& ws) { bool threaded = ws.GetThreadPool().NumThreads() > 1; for (int output_idx = 0; output_idx < num_outputs; output_idx++) { - auto& output = ws.OutputRef(output_idx); + auto& output = ws.Output(output_idx); for (int data_idx = 0; data_idx < num_samples; data_idx++) { auto& sample = GetSample(data_idx); ThreadPool::Work copy_task = [output_idx = output_idx, data_idx = data_idx, &output, @@ -73,20 +73,20 @@ DALI_SCHEMA(readers__Webdataset) The webdataset format is a way of providing efficient access to datasets stored in tar archives. -Storing data in POSIX tar archives greatly speeds up I/O operations on mechanical storage devices -and on network file systems because it allows the operating system to reduce the number of I/O +Storing data in POSIX tar archives greatly speeds up I/O operations on mechanical storage devices +and on network file systems because it allows the operating system to reduce the number of I/O operations and to read the data ahead. -WebDataset fulfils a similar function to Tensorflow's TFRecord/tf.Example classes, but is much -easier to adopt because it does not actually require any data conversion. The data is stored in -exactly the same format inside tar files as it is on disk, and all preprocessing and data +WebDataset fulfils a similar function to Tensorflow's TFRecord/tf.Example classes, but is much +easier to adopt because it does not actually require any data conversion. The data is stored in +exactly the same format inside tar files as it is on disk, and all preprocessing and data augmentation code remains unchanged. -The dataset consists of one or more tar archives, each of which is further split into samples. -A sample contains one or more components that correspond to the actual files contained within -the archive. The components that belong to a specific sample are aggregated by filename without +The dataset consists of one or more tar archives, each of which is further split into samples. +A sample contains one or more components that correspond to the actual files contained within +the archive. The components that belong to a specific sample are aggregated by filename without extension (for the specifics about the extensions please read the description of the ``ext`` parameter -below). Note that samples with their filename starting with a dot will not be loaded, as well as +below). Note that samples with their filename starting with a dot will not be loaded, as well as entries that are not regular files. In addition to the tar archive with data, each archive should come with a corresponding index file. @@ -117,7 +117,7 @@ Has to be the same length as the ``index_paths`` argument.)code", .AddArg("ext", R"code(The extension sets for each of the outputs produced. The number of extension sets determines the number of outputs of the reader. -The extensions of the components are counted as the text after the first dot in the name of the file +The extensions of the components are counted as the text after the first dot in the name of the file (excluding the samples starting with a dot). The different extension options should be separated with a semicolon (';') and may contain dots. @@ -125,7 +125,7 @@ Example: "left.png;right.jpg")code", DALI_STRING_VEC) .AddOptionalArg("index_paths", R"code(The list of the index files corresponding to the respective webdataset archives. - + Has to be the same length as the ``paths`` argument. In case it is not provided, it will be inferred automatically from the webdataset archive.)code", std::vector()) diff --git a/dali/operators/segmentation/random_mask_pixel.cc b/dali/operators/segmentation/random_mask_pixel.cc index 223ce57faf2..43c30b455c2 100644 --- a/dali/operators/segmentation/random_mask_pixel.cc +++ b/dali/operators/segmentation/random_mask_pixel.cc @@ -88,7 +88,7 @@ RandomMaskPixelCPU::RandomMaskPixelCPU(const OpSpec &spec) bool RandomMaskPixelCPU::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &in_masks = ws.template InputRef(0); + const auto &in_masks = ws.template Input(0); int nsamples = in_masks.num_samples(); auto in_masks_shape = in_masks.shape(); int ndim = in_masks_shape.sample_dim(); @@ -111,8 +111,8 @@ bool RandomMaskPixelCPU::SetupImpl(std::vector &output_desc, template void RandomMaskPixelCPU::RunImplTyped(workspace_t &ws) { - const auto &in_masks = ws.template InputRef(0); - auto &out_pixel_pos = ws.template OutputRef(0); + const auto &in_masks = ws.template Input(0); + auto &out_pixel_pos = ws.template Output(0); int nsamples = in_masks.num_samples(); auto in_masks_shape = in_masks.shape(); int ndim = in_masks_shape.sample_dim(); @@ -171,7 +171,7 @@ void RandomMaskPixelCPU::RunImplTyped(workspace_t &ws) { } void RandomMaskPixelCPU::RunImpl(workspace_t &ws) { - const auto &in_masks = ws.template InputRef(0); + const auto &in_masks = ws.template Input(0); TYPE_SWITCH(in_masks.type(), type2id, T, MASK_SUPPORTED_TYPES, ( RunImplTyped(ws); ), ( // NOLINT diff --git a/dali/operators/segmentation/random_object_bbox.cc b/dali/operators/segmentation/random_object_bbox.cc index 1a494c26122..322748bc6c3 100644 --- a/dali/operators/segmentation/random_object_bbox.cc +++ b/dali/operators/segmentation/random_object_bbox.cc @@ -113,7 +113,7 @@ to compute than to recalculate the object boxes.)", false); bool RandomObjectBBox::SetupImpl(vector &out_descs, const HostWorkspace &ws) { out_descs.resize(spec_.NumOutput()); - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); int ndim = input.sample_dim(); int N = input.num_samples(); @@ -552,7 +552,7 @@ void RandomObjectBBox::AllocateTempStorage(const TensorVector &input } void RandomObjectBBox::RunImpl(HostWorkspace &ws) { - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); int N = input.num_samples(); if (N == 0) return; @@ -560,13 +560,13 @@ void RandomObjectBBox::RunImpl(HostWorkspace &ws) { int ndim = input.sample_dim(); auto &tp = ws.GetThreadPool(); - OutListCPU out1 = view(ws.OutputRef(0)); + OutListCPU out1 = view(ws.Output(0)); OutListCPU out2; if (format_ != Out_Box) - out2 = view(ws.OutputRef(1)); + out2 = view(ws.Output(1)); OutListCPU class_label_out; if (HasClassLabelOutput()) - class_label_out = view(ws.OutputRef(class_output_idx_)); + class_label_out = view(ws.Output(class_output_idx_)); TensorShape<> default_anchor; default_anchor.resize(ndim); diff --git a/dali/operators/segmentation/select_masks.cc b/dali/operators/segmentation/select_masks.cc index dcd0d9fa643..96617854ee2 100644 --- a/dali/operators/segmentation/select_masks.cc +++ b/dali/operators/segmentation/select_masks.cc @@ -87,12 +87,12 @@ in ``mask_ids`` input.)code", bool SelectMasksCPU::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &in_mask_ids = ws.template InputRef(0); + const auto &in_mask_ids = ws.template Input(0); auto in_mask_ids_shape = in_mask_ids.shape(); DALI_ENFORCE(in_mask_ids.type() == DALI_INT32, "``mask_ids`` input is expected to be int32"); DALI_ENFORCE(in_mask_ids_shape.sample_dim() == 1, "``mask_ids`` input is expected to be 1D"); - const auto &in_polygons = ws.template InputRef(1); + const auto &in_polygons = ws.template Input(1); auto in_polygons_shape = in_polygons.shape(); DALI_ENFORCE(in_polygons.type() == DALI_INT32, "``polygons`` input is expected to be int32"); @@ -100,7 +100,7 @@ bool SelectMasksCPU::SetupImpl(std::vector &output_desc, make_string("``polygons`` input is expected to be 2D. Got ", in_polygons_shape.sample_dim(), "D")); - const auto &in_vertices = ws.template InputRef(2); + const auto &in_vertices = ws.template Input(2); auto in_vertices_shape = in_vertices.shape(); DALI_ENFORCE(in_vertices_shape.sample_dim() == 2, make_string("``vertices`` input is expected to be 2D. Got ", @@ -194,14 +194,14 @@ bool SelectMasksCPU::SetupImpl(std::vector &output_desc, template void SelectMasksCPU::RunImplTyped(workspace_t &ws) { // Inputs were already validated and input 0 was already parsed in SetupImpl - const auto &in_polygons = ws.template InputRef(1); + const auto &in_polygons = ws.template Input(1); const auto &in_polygons_view = view(in_polygons); - auto &out_polygons = ws.template OutputRef(0); + auto &out_polygons = ws.template Output(0); const auto &out_polygons_view = view(out_polygons); - const auto &in_vertices = ws.template InputRef(2); + const auto &in_vertices = ws.template Input(2); const auto &in_vertices_view = reinterpret_view(in_vertices); - auto &out_vertices = ws.template OutputRef(1); + auto &out_vertices = ws.template Output(1); const auto &out_vertices_view = reinterpret_view(out_vertices); for (int i = 0; i < in_polygons_view.num_samples(); i++) { @@ -231,7 +231,7 @@ void SelectMasksCPU::RunImplTyped(workspace_t &ws) { } void SelectMasksCPU::RunImpl(workspace_t &ws) { - const auto &in_vertices = ws.template InputRef(2); + const auto &in_vertices = ws.template Input(2); VALUE_SWITCH(in_vertices.type_info().size(), dtype_sz, (1, 2, 4, 8, 16), ( using T = kernels::type_of_size; RunImplTyped(ws); diff --git a/dali/operators/sequence/element_extract.h b/dali/operators/sequence/element_extract.h index d847998ce5e..7aaf4f891b9 100644 --- a/dali/operators/sequence/element_extract.h +++ b/dali/operators/sequence/element_extract.h @@ -87,7 +87,7 @@ class ElementExtract : public Operator { } bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc.resize(element_map_.size()); auto output_shape = detail::GetOutputShape(input.shape(), element_map_, input.GetLayout()); for (auto &desc : output_desc) { @@ -98,13 +98,13 @@ class ElementExtract : public Operator { } void RunImpl(workspace_t &ws) override { - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); auto element_layout = VideoLayoutInfo::GetFrameLayout(input.GetLayout()); int elements_per_sample = element_map_.size(); auto data_type = input.type_info(); for (int k = 0; k < elements_per_sample; k++) { int element = element_map_[k]; - auto &output = ws.template OutputRef(k); + auto &output = ws.template Output(k); for (unsigned int i = 0; i < input.num_samples(); i++) { auto tensor_shape = input.tensor_shape(i); auto element_size = volume(tensor_shape.begin() + 1, tensor_shape.end()); diff --git a/dali/operators/sequence/optical_flow/optical_flow.cc b/dali/operators/sequence/optical_flow/optical_flow.cc index aeb88c73169..d418e72f542 100644 --- a/dali/operators/sequence/optical_flow/optical_flow.cc +++ b/dali/operators/sequence/optical_flow/optical_flow.cc @@ -113,9 +113,9 @@ void OpticalFlow::RunImpl(Workspace &ws) { if (enable_external_hints_) { // Fetch data // Input is a TensorList, where every Tensor is a sequence - const auto &input = ws.InputRef(0); - const auto &hints = ws.InputRef(1); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + const auto &hints = ws.Input(1); + auto &output = ws.Output(0); output.SetLayout("HWC"); // Channels represent the two flow vector components (x and y) // Extract calculation params ExtractParams(input, hints); @@ -153,8 +153,8 @@ void OpticalFlow::RunImpl(Workspace &ws) { } else { // Fetch data // Input is a TensorList, where every Tensor is a sequence - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); output.SetLayout(input.GetLayout()); // Extract calculation params diff --git a/dali/operators/sequence/sequence_rearrange.cc b/dali/operators/sequence/sequence_rearrange.cc index 9e6d943afab..70d4a05b654 100644 --- a/dali/operators/sequence/sequence_rearrange.cc +++ b/dali/operators/sequence/sequence_rearrange.cc @@ -70,8 +70,8 @@ copy_desc GetCopyDesc(char *output_sample, const char *input_sample, int out_ele template <> void SequenceRearrange::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto &thread_pool = ws.GetThreadPool(); auto out_shape = output.shape(); auto curr_batch_size = ws.GetInputBatchSize(0); diff --git a/dali/operators/sequence/sequence_rearrange.cu b/dali/operators/sequence/sequence_rearrange.cu index 8cca7a84802..359d4a78a05 100644 --- a/dali/operators/sequence/sequence_rearrange.cu +++ b/dali/operators/sequence/sequence_rearrange.cu @@ -23,8 +23,8 @@ namespace dali { template <> void SequenceRearrange::RunImpl(workspace_t &ws) { scatter_gather_.Reset(); - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); const TypeInfo &type = input.type_info(); auto curr_batch_size = ws.GetInputBatchSize(0); diff --git a/dali/operators/sequence/sequence_rearrange.h b/dali/operators/sequence/sequence_rearrange.h index 5fb462ea5b4..865d862d451 100644 --- a/dali/operators/sequence/sequence_rearrange.h +++ b/dali/operators/sequence/sequence_rearrange.h @@ -1,4 +1,4 @@ -// Copyright (c) 2019-2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2019-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -66,7 +66,7 @@ class SequenceRearrange : public Operator { } bool SetupImpl(std::vector& output_desc, const workspace_t& ws) override { - const auto& input = ws.template InputRef(0); + const auto& input = ws.template Input(0); const auto& in_shape = input.shape(); // temporary in some cases DALI_ENFORCE(in_shape.sample_dim() > 1, "Sequence elements must have at least 1 dim"); output_desc.resize(1); diff --git a/dali/operators/signal/decibel/to_decibels_op_cpu.cc b/dali/operators/signal/decibel/to_decibels_op_cpu.cc index c42e2e54f22..29f0a466644 100644 --- a/dali/operators/signal/decibel/to_decibels_op_cpu.cc +++ b/dali/operators/signal/decibel/to_decibels_op_cpu.cc @@ -51,8 +51,8 @@ template <> bool ToDecibels::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(kNumOutputs); - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); kernels::KernelContext ctx; auto in_shape = input.shape(); int nsamples = input.num_samples(); @@ -74,8 +74,8 @@ bool ToDecibels::SetupImpl(std::vector &output_desc, template <> void ToDecibels::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_shape = input.shape(); auto& thread_pool = ws.GetThreadPool(); diff --git a/dali/operators/signal/decibel/to_decibels_op_gpu.cu b/dali/operators/signal/decibel/to_decibels_op_gpu.cu index 5741041f1cc..d873d720949 100644 --- a/dali/operators/signal/decibel/to_decibels_op_gpu.cu +++ b/dali/operators/signal/decibel/to_decibels_op_gpu.cu @@ -57,7 +57,7 @@ class ToDecibelsImpl : public OpImplBase { template bool ToDecibelsImpl::SetupImpl(std::vector &output_desc, const workspace_t &ws) { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); auto in_view = view(input); auto type = type2id::value; @@ -80,8 +80,8 @@ bool ToDecibelsImpl::SetupImpl(std::vector &output_desc, template void ToDecibelsImpl::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_view = view(input); auto out_view = view(output); @@ -103,7 +103,7 @@ template <> bool ToDecibels::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(kNumOutputs); - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); auto type = input.type(); TYPE_SWITCH(type, type2id, T, (float), ( using Impl = ToDecibelsImpl; diff --git a/dali/operators/signal/fft/power_spectrum.cc b/dali/operators/signal/fft/power_spectrum.cc index a280bd69963..75b311f5746 100644 --- a/dali/operators/signal/fft/power_spectrum.cc +++ b/dali/operators/signal/fft/power_spectrum.cc @@ -58,8 +58,8 @@ template <> bool PowerSpectrum::SetupImpl(std::vector &output_desc, const workspace_t &ws) { output_desc.resize(kNumOutputs); - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); kernels::KernelContext ctx; auto in_shape = input.shape(); int nsamples = input.num_samples(); @@ -86,8 +86,8 @@ bool PowerSpectrum::SetupImpl(std::vector &output_desc, template <> void PowerSpectrum::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto in_shape = input.shape(); auto& thread_pool = ws.GetThreadPool(); // Other types not supported for now diff --git a/dali/operators/signal/fft/spectrogram.cc b/dali/operators/signal/fft/spectrogram.cc index 033dde11fc6..c2e9404c7b7 100644 --- a/dali/operators/signal/fft/spectrogram.cc +++ b/dali/operators/signal/fft/spectrogram.cc @@ -179,8 +179,8 @@ SpectrogramImplCpu::SpectrogramImplCpu(const OpSpec &spec) template bool SpectrogramImplCpu::SetupImpl(std::vector &out_desc, const workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); kernels::KernelContext ctx; auto in_shape = input.shape(); int nsamples = input.num_samples(); @@ -254,8 +254,8 @@ bool SpectrogramImplCpu::SetupImpl(std::vector &out_desc template void SpectrogramImplCpu::RunImpl(workspace_t &ws) { - const auto &input = ws.InputRef(0); - auto &output = ws.OutputRef(0); + const auto &input = ws.Input(0); + auto &output = ws.Output(0); auto out_shape = output.shape(); int nsamples = input.num_samples(); auto& thread_pool = ws.GetThreadPool(); diff --git a/dali/operators/signal/fft/spectrogram_gpu.cc b/dali/operators/signal/fft/spectrogram_gpu.cc index ee14a91962b..f892ae1e158 100644 --- a/dali/operators/signal/fft/spectrogram_gpu.cc +++ b/dali/operators/signal/fft/spectrogram_gpu.cc @@ -75,7 +75,7 @@ struct SpectrogramOpImplGPU : public OpImplBase { } bool SetupImpl(std::vector &output_desc, const DeviceWorkspace &ws) override { - auto &in = ws.InputRef(0); + auto &in = ws.Input(0); KernelContext ctx; ctx.gpu.stream = ws.stream(); const auto &in_shape = in.shape(); @@ -134,8 +134,8 @@ struct SpectrogramOpImplGPU : public OpImplBase { } void RunImpl(DeviceWorkspace &ws) override { - const auto &in = ws.InputRef(0); - auto &out = ws.OutputRef(0); + const auto &in = ws.Input(0); + auto &out = ws.Output(0); out.SetLayout(layout); auto kernel_output_shape = kmgr.GetRequirements(0).output_shapes[0].to_static<2>(); const auto in_view_1D = reshape(view(in), in_shape_1D); diff --git a/dali/operators/ssd/box_encoder.cc b/dali/operators/ssd/box_encoder.cc index cf18f70b776..caa37fc6aa0 100644 --- a/dali/operators/ssd/box_encoder.cc +++ b/dali/operators/ssd/box_encoder.cc @@ -156,8 +156,8 @@ void BoxEncoder::WriteMatchesToOutput( } void BoxEncoder::RunImpl(SampleWorkspace &ws) { - const auto &bboxes_input = ws.InputRef(kBoxesInId); - const auto &labels_input = ws.InputRef(kLabelsInId); + const auto &bboxes_input = ws.Input(kBoxesInId); + const auto &labels_input = ws.Input(kLabelsInId); const auto num_boxes = bboxes_input.dim(0); const auto labels = labels_input.data(); @@ -166,11 +166,11 @@ void BoxEncoder::RunImpl(SampleWorkspace &ws) { ReadBoxes(make_span(boxes), make_cspan(bboxes_input.data(), bboxes_input.size()), {}, {}); // Create output - auto &bboxes_output = ws.OutputRef(kBoxesOutId); + auto &bboxes_output = ws.Output(kBoxesOutId); bboxes_output.Resize({static_cast(anchors_.size()), BoundingBox::size}, bboxes_input.type()); auto out_boxes = bboxes_output.mutable_data(); - auto &labels_output = ws.OutputRef(kLabelsOutId); + auto &labels_output = ws.Output(kLabelsOutId); labels_output.Resize({static_cast(anchors_.size())}, labels_input.type()); auto out_labels = labels_output.mutable_data(); diff --git a/dali/operators/ssd/box_encoder.cu b/dali/operators/ssd/box_encoder.cu index 594c058915a..cde92d853a8 100644 --- a/dali/operators/ssd/box_encoder.cu +++ b/dali/operators/ssd/box_encoder.cu @@ -260,8 +260,8 @@ BoxEncoder::CalculateDims( } void BoxEncoder::RunImpl(Workspace &ws) { - const auto &boxes_input = ws.InputRef(kBoxesInId); - const auto &labels_input = ws.InputRef(kLabelsInId); + const auto &boxes_input = ws.Input(kBoxesInId); + const auto &labels_input = ws.Input(kLabelsInId); assert(ws.GetInputBatchSize(kBoxesInId) == ws.GetInputBatchSize(kLabelsInId)); auto curr_batch_size = ws.GetInputBatchSize(kBoxesInId); @@ -273,10 +273,10 @@ void BoxEncoder::RunImpl(Workspace &ws) { auto dims = CalculateDims(boxes_input); - auto &boxes_output = ws.OutputRef(kBoxesOutId); + auto &boxes_output = ws.Output(kBoxesOutId); boxes_output.Resize(dims.first, boxes_input.type()); - auto &labels_output = ws.OutputRef(kLabelsOutId); + auto &labels_output = ws.Output(kLabelsOutId); labels_output.Resize(dims.second, labels_input.type()); samples.resize(curr_batch_size_); diff --git a/dali/operators/ssd/box_encoder_test.cc b/dali/operators/ssd/box_encoder_test.cc index 468da270afe..ceecc477e4f 100644 --- a/dali/operators/ssd/box_encoder_test.cc +++ b/dali/operators/ssd/box_encoder_test.cc @@ -1085,14 +1085,14 @@ class BoxEncoderTest : public GenericBBoxesTest { } void CheckAnswersForCocoOnCpu(DeviceWorkspace *ws, bool offset = false) { - TensorList &boxes = ws->OutputRef(0); - TensorList &labels = ws->OutputRef(1); + TensorList &boxes = ws->Output(0); + TensorList &labels = ws->Output(1); CheckAnswersForCoco(&boxes, &labels, offset); } void CheckAnswersForCocoOnGpu(DeviceWorkspace *ws, bool offset = false) { - auto boxes = this->CopyTensorListToHost(ws->OutputRef(0)); - auto labels = this->CopyTensorListToHost(ws->OutputRef(1)); + auto boxes = this->CopyTensorListToHost(ws->Output(0)); + auto labels = this->CopyTensorListToHost(ws->Output(1)); CheckAnswersForCoco(boxes.get(), labels.get(), offset); } }; diff --git a/dali/operators/ssd/random_crop.cc b/dali/operators/ssd/random_crop.cc index 781313baeb8..d007718c4b5 100644 --- a/dali/operators/ssd/random_crop.cc +++ b/dali/operators/ssd/random_crop.cc @@ -179,10 +179,10 @@ void crop(const Tensor& img, vector bounds, Tensor& template <> void SSDRandomCrop::RunImpl(SampleWorkspace &ws) { // [H, W, C], dtype=uint8_t - const auto& img = ws.InputRef(0); + const auto& img = ws.Input(0); // [N] : [ltrb, ... ], dtype=float - const auto& bboxes = ws.InputRef(1); - const auto& labels = ws.InputRef(2); + const auto& bboxes = ws.Input(1); + const auto& labels = ws.Input(2); int sample = ws.data_idx(); auto N = bboxes.dim(0); @@ -202,9 +202,9 @@ void SSDRandomCrop::RunImpl(SampleWorkspace &ws) { if (option.no_crop()) { // copy directly to output without modification - ws.OutputRef(0).Copy(img, 0); - ws.OutputRef(1).Copy(bboxes, 0); - ws.OutputRef(2).Copy(labels, 0); + ws.Output(0).Copy(img, 0); + ws.Output(1).Copy(bboxes, 0); + ws.Output(2).Copy(labels, 0); return; } @@ -273,10 +273,10 @@ void SSDRandomCrop::RunImpl(SampleWorkspace &ws) { // now we know how many output bboxes there will be, we can allocate // the output. - auto &img_out = ws.OutputRef(0); + auto &img_out = ws.Output(0); img_out.SetLayout(img.GetLayout()); - auto &bbox_out = ws.OutputRef(1); - auto &label_out = ws.OutputRef(2); + auto &bbox_out = ws.Output(1); + auto &label_out = ws.Output(2); bbox_out.Resize({valid_bboxes, 4}, DALI_FLOAT); auto *bbox_out_data = bbox_out.mutable_data(); diff --git a/dali/pipeline/executor/executor.cc b/dali/pipeline/executor/executor.cc index 0946f5a1847..6203c524dd7 100644 --- a/dali/pipeline/executor/executor.cc +++ b/dali/pipeline/executor/executor.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2020-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -301,9 +301,9 @@ void Executor::RunHelper(OpNode &op_node, Workspac for (int i = 0; i < spec.NumRegularInput(); i++) { bool had_empty_layout = false; if (ws.template InputIsType(i)) { - had_empty_layout = SetDefaultLayoutIfNeeded(ws.template InputRef(i), schema, i); + had_empty_layout = SetDefaultLayoutIfNeeded(ws.template Input(i), schema, i); } else { - had_empty_layout = SetDefaultLayoutIfNeeded(ws.template InputRef(i), schema, i); + had_empty_layout = SetDefaultLayoutIfNeeded(ws.template Input(i), schema, i); } if (had_empty_layout) empty_layout_in_idxs.push_back(i); } @@ -318,9 +318,9 @@ void Executor::RunHelper(OpNode &op_node, Workspac for (int i = 0; i < ws.NumOutput(); i++) { auto &desc = output_desc[i]; if (ws.template OutputIsType(i)) { - ws.template OutputRef(i).Resize(desc.shape, desc.type); + ws.template Output(i).Resize(desc.shape, desc.type); } else { - ws.template OutputRef(i).Resize(desc.shape, desc.type); + ws.template Output(i).Resize(desc.shape, desc.type); } } } else { @@ -334,10 +334,10 @@ void Executor::RunHelper(OpNode &op_node, Workspac for (int i : empty_layout_in_idxs) { if (ws.template InputIsType(i)) { - auto &in = ws.template InputRef(i); + auto &in = ws.template Input(i); in.SetLayout({}); } else { - auto &in = ws.template InputRef(i); + auto &in = ws.template Input(i); in.SetLayout({}); } } diff --git a/dali/pipeline/executor/executor.h b/dali/pipeline/executor/executor.h index 3a52e512bed..7c11a5d6dc7 100644 --- a/dali/pipeline/executor/executor.h +++ b/dali/pipeline/executor/executor.h @@ -195,12 +195,12 @@ class DLL_PUBLIC Executor : public ExecutorBase, public WorkspacePolicy, public reserved_size = 0; max_reserved_size = 0; if (ws.template OutputIsType(i)) { - auto &out = ws.template OutputRef(i); + auto &out = ws.template Output(i); out_size = out.nbytes(); reserved_size = out.capacity(); GetMaxSizes(out, max_out_size, max_reserved_size); } else { - auto &out = ws.template OutputRef(i); + auto &out = ws.template Output(i); out_size = out.nbytes(); reserved_size = out.capacity(); GetMaxSizes(out, max_out_size, max_reserved_size); diff --git a/dali/pipeline/executor/executor_test.cc b/dali/pipeline/executor/executor_test.cc index 2b724ba9e1e..a64f79fa655 100644 --- a/dali/pipeline/executor/executor_test.cc +++ b/dali/pipeline/executor/executor_test.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -599,7 +599,7 @@ TYPED_TEST(ExecutorSyncTest, TestPrefetchedExecution) { ASSERT_EQ(ws.NumOutput(), 1); ASSERT_EQ(ws.NumInput(), 0); ASSERT_TRUE(ws.OutputIsType(0)); - TensorList &res1 = ws.OutputRef(0); + TensorList &res1 = ws.Output(0); for (int i = 0; i < batch_size; ++i) { this->VerifyDecode( res1.template tensor(i), @@ -615,7 +615,7 @@ TYPED_TEST(ExecutorSyncTest, TestPrefetchedExecution) { auto status_2 = barrier_future_2.wait_for(std::chrono::seconds(5)); ASSERT_EQ(status_2, std::future_status::ready); ASSERT_EQ(cb_counter, 2); - TensorList &res2 = ws.OutputRef(0); + TensorList &res2 = ws.Output(0); for (int i = 0; i < batch_size; ++i) { this->VerifyDecode( res2.template tensor(i), diff --git a/dali/pipeline/operator/builtin/copy.h b/dali/pipeline/operator/builtin/copy.h index fa3a43d8b0a..23c5759ea33 100644 --- a/dali/pipeline/operator/builtin/copy.h +++ b/dali/pipeline/operator/builtin/copy.h @@ -41,16 +41,16 @@ class Copy : public Operator { bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { output_desc.resize(1); - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc[0].type = input.type(); output_desc[0].shape = input.shape(); return true; } void RunImpl(workspace_t &ws) override { - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); auto data_type_size = input.type_info().size(); - auto &output = ws.template OutputRef(0); + auto &output = ws.template Output(0); output.SetLayout(input.GetLayout()); for (unsigned int i = 0; i < input.num_samples(); i++) { auto tensor_shape = input.tensor_shape(i); diff --git a/dali/pipeline/operator/builtin/external_source.cc b/dali/pipeline/operator/builtin/external_source.cc index 0e43e08c1bb..20cc6f4cbe4 100644 --- a/dali/pipeline/operator/builtin/external_source.cc +++ b/dali/pipeline/operator/builtin/external_source.cc @@ -25,7 +25,7 @@ void ExternalSource::RunImpl(HostWorkspace &ws) { tensor_vector_elm = tv_data_.PopFront(); state_.pop_front(); } - auto &output = ws.template OutputRef(0); + auto &output = ws.template Output(0); // if the output is pinned and input not it needs to be copied if (output.is_pinned() && !tensor_vector_elm.front()->is_pinned()) { auto &thread_pool = ws.GetThreadPool(); @@ -36,7 +36,7 @@ void ExternalSource::RunImpl(HostWorkspace &ws) { for (int sample_id = 0; sample_id < curr_batch_size; ++sample_id) { thread_pool.AddWork( [&ws, sample_id, &tensor_vector_elm](int tid) { - Tensor &output_tensor = ws.OutputRef(0)[sample_id]; + Tensor &output_tensor = ws.Output(0)[sample_id]; // HostWorkspace doesn't have any stream cudaStream_t stream = 0; output_tensor.Copy((*tensor_vector_elm.front())[sample_id], stream); @@ -46,7 +46,7 @@ void ExternalSource::RunImpl(HostWorkspace &ws) { thread_pool.RunAll(); // as we copy element by element and the output is contiguous we need to set layout // for the whole output not each element(view) - auto &output = ws.template OutputRef(0); + auto &output = ws.template Output(0); output.SetLayout(tensor_vector_elm.front()->GetLayout()); } else { // swap output with tensor_vector_elm content diff --git a/dali/pipeline/operator/builtin/external_source.cu b/dali/pipeline/operator/builtin/external_source.cu index 53dd0046fd6..ddcdae2c6d0 100644 --- a/dali/pipeline/operator/builtin/external_source.cu +++ b/dali/pipeline/operator/builtin/external_source.cu @@ -37,7 +37,7 @@ void ExternalSource::RunImpl(DeviceWorkspace &ws) { } } - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); cudaStream_t stream_used = ws.has_stream() ? ws.stream() : 0; if (!state_info.no_copy || state_info.copied_shared_data) { CUDA_CALL(cudaStreamWaitEvent(stream_used, *internal_copy_to_storage.front(), 0)); diff --git a/dali/pipeline/operator/builtin/external_source_test.cc b/dali/pipeline/operator/builtin/external_source_test.cc index eb6d5c5aff0..d7b9ba55cf9 100644 --- a/dali/pipeline/operator/builtin/external_source_test.cc +++ b/dali/pipeline/operator/builtin/external_source_test.cc @@ -192,7 +192,7 @@ class ExternalSourceTest : public::testing::WithParamInterface, bool RunOutputs() { DeviceWorkspace ws; exe_->Outputs(&ws); - auto &tensor_gpu_list = ws.OutputRef(0); + auto &tensor_gpu_list = ws.Output(0); TensorList tensor_cpu_list; tensor_cpu_list.Copy(tensor_gpu_list, (ws.has_stream() ? ws.stream() : 0)); CUDA_CALL(cudaStreamSynchronize(ws.has_stream() ? ws.stream() : 0)); @@ -579,9 +579,9 @@ void TestRunExternalSource(Pipeline &pipe, const std::string &name, TensorList output_cpu; pipe.Outputs(&ws); if (dev == "cpu") { - output_cpu.Copy(ws.OutputRef(0), 0); + output_cpu.Copy(ws.Output(0), 0); } else { - output_cpu.Copy(ws.OutputRef(0), 0); + output_cpu.Copy(ws.Output(0), 0); cudaStreamSynchronize(0); } ASSERT_EQ(input_cpu.shape(), output_cpu.shape()); diff --git a/dali/pipeline/operator/builtin/make_contiguous.cc b/dali/pipeline/operator/builtin/make_contiguous.cc index 0ffb36fc148..605d6c6131d 100644 --- a/dali/pipeline/operator/builtin/make_contiguous.cc +++ b/dali/pipeline/operator/builtin/make_contiguous.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -17,8 +17,8 @@ namespace dali { void MakeContiguousCPU::RunImpl(HostWorkspace &ws) { - auto &input = ws.template InputRef(0); - auto &output = ws.template OutputRef(0); + auto &input = ws.template Input(0); + auto &output = ws.template Output(0); int batch_size = input.num_samples(); output.SetLayout(input.GetLayout()); auto shapes = input.shape(); diff --git a/dali/pipeline/operator/builtin/make_contiguous.cu b/dali/pipeline/operator/builtin/make_contiguous.cu index 27aa9b8fa92..8e3c98eccbd 100644 --- a/dali/pipeline/operator/builtin/make_contiguous.cu +++ b/dali/pipeline/operator/builtin/make_contiguous.cu @@ -18,13 +18,13 @@ namespace dali { void MakeContiguousMixed::Run(MixedWorkspace &ws) { - const auto& input = ws.template InputRef(0); + const auto& input = ws.template Input(0); int sample_dim = input[0].shape().sample_dim(); size_t batch_size = input.num_samples(); DALIDataType type = input.type(); for (size_t i = 0; i < input.num_samples(); ++i) { - auto &sample = ws.InputRef(0)[i]; + auto &sample = ws.Input(0)[i]; size_t sample_bytes = sample.nbytes(); if (coalesced && sample_bytes > COALESCE_THRESHOLD) coalesced = false; @@ -34,7 +34,7 @@ void MakeContiguousMixed::Run(MixedWorkspace &ws) { "in input batch. Cannot copy to contiguous device buffer."); } - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); if (coalesced) { DomainTimeRange tr("[DALI][MakeContiguousMixed] coalesced", DomainTimeRange::kBlue); cpu_output_buff.Copy(input, 0); diff --git a/dali/pipeline/operator/builtin/make_contiguous.h b/dali/pipeline/operator/builtin/make_contiguous.h index c69ac708c57..9b2e572b65b 100644 --- a/dali/pipeline/operator/builtin/make_contiguous.h +++ b/dali/pipeline/operator/builtin/make_contiguous.h @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2020, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -47,7 +47,7 @@ class MakeContiguousBase : public Operator { bool SetupImpl(std::vector &output_desc, const workspace_t &ws) override { output_desc.resize(1); - auto &input = ws.template InputRef(0); + auto &input = ws.template Input(0); output_desc[0].shape = input.shape(); output_desc[0].type = input.type(); return true; diff --git a/dali/pipeline/operator/op_spec_test.cc b/dali/pipeline/operator/op_spec_test.cc index 183633a5ae9..6fd0c9a410d 100644 --- a/dali/pipeline/operator/op_spec_test.cc +++ b/dali/pipeline/operator/op_spec_test.cc @@ -323,17 +323,17 @@ class TestArgumentInput_Producer : public Operator { void RunImpl(HostWorkspace &ws) override { // Initialize all the data with a 0, 1, 2 .... sequence - auto &out0 = ws.OutputRef(0); + auto &out0 = ws.Output(0); for (int i = 0; i < out0.shape().num_samples(); i++) { *out0[i].mutable_data() = i; } - auto &out1 = ws.OutputRef(1); + auto &out1 = ws.Output(1); for (int i = 0; i < out1.shape().num_samples(); i++) { *out1[i].mutable_data() = i; } - auto &out2 = ws.OutputRef(2); + auto &out2 = ws.Output(2); for (int i = 0; i < out2.shape().num_samples(); i++) { for (int j = 0; j < 2; j++) { out2[i].mutable_data()[j] = i; diff --git a/dali/pipeline/operator/operator.cc b/dali/pipeline/operator/operator.cc index 1f5be4259df..ff10689e995 100644 --- a/dali/pipeline/operator/operator.cc +++ b/dali/pipeline/operator/operator.cc @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. // // Licensed under the Apache License, Version 2.0 (the "License"); // you may not use this file except in compliance with the License. @@ -36,7 +36,7 @@ template void OperatorBase::EnforceUniformOutputBatchSize(const workspace_t &ws) const { auto ref_batch_size = ws.NumInput() > 0 ? ws.GetInputBatchSize(0) : ws.GetRequestedBatchSize(0); for (int i = 0; i < ws.NumOutput(); i++) { - auto output_batch_size = ws.template OutputRef(i).shape().num_samples(); + auto output_batch_size = ws.template Output(i).shape().num_samples(); DALI_ENFORCE(ref_batch_size == output_batch_size, make_string("Batch size has to be uniform across one iteration. Expected: ", ref_batch_size, "; Actual: ", output_batch_size)); @@ -50,7 +50,7 @@ void OperatorBase::EnforceUniformOutputBatchSize( auto ref_batch_size = ws.NumInput() > 0 ? ws.GetInputBatchSize(0) : ws.GetRequestedBatchSize(0); for (int i = 0; i < ws.NumOutput(); i++) { auto output_batch_size = const_cast &>(ws) - .template OutputRef(i) + .template Output(i) .shape() .num_samples(); DALI_ENFORCE(ref_batch_size == output_batch_size, diff --git a/dali/pipeline/operator/operator.h b/dali/pipeline/operator/operator.h index f695e8f5a63..2e82760b4d5 100644 --- a/dali/pipeline/operator/operator.h +++ b/dali/pipeline/operator/operator.h @@ -70,10 +70,10 @@ inline void CheckInputLayouts(const Workspace &ws, const OpSpec &spec) { auto &schema = spec.GetSchema(); for (int i = 0; i < spec.NumRegularInput(); ++i) { if (ws.template InputIsType(i)) { - auto &input = ws.template InputRef(i); + auto &input = ws.template Input(i); (void) schema.GetInputLayout(i, input.shape().sample_dim(), input.GetLayout()); } else if (ws.template InputIsType(i)) { - auto &input = ws.template InputRef(i); + auto &input = ws.template Input(i); (void) schema.GetInputLayout(i, input.shape().sample_dim(), input.GetLayout()); } else { DALI_FAIL(make_string("Input ", i, " has an unknown backend")); @@ -255,24 +255,24 @@ class Operator : public OperatorBase {}; template TensorLayout GetInputLayout(Workspace& ws, int i) { if (ws.template InputIsType(i)) { - auto &in = ws.template InputRef(i); + auto &in = ws.template Input(i); return in.GetLayout(); } assert(ws.template InputIsType(i)); - auto &in = ws.template InputRef(i); + auto &in = ws.template Input(i); return in.GetLayout(); } template TensorLayout GetOutputLayout(Workspace &ws, int i) { if (ws.template OutputIsType(i)) { - auto &out = ws.template OutputRef(i); + auto &out = ws.template Output(i); return out.GetLayout(); } assert(ws.template OutputIsType(i)); - auto &out = ws.template OutputRef(i); + auto &out = ws.template Output(i); return out.GetLayout(); } @@ -325,7 +325,7 @@ class Operator : public OperatorBase { auto curr_batch_size = ws.NumInput() > 0 ? ws.GetInputBatchSize(0) : max_batch_size_; for (int i = 0; i < ws.NumOutput(); i++) { - auto &output = ws.OutputRef(i); + auto &output = ws.Output(i); output.SetSize(curr_batch_size); } auto &thread_pool = ws.GetThreadPool(); diff --git a/dali/pipeline/pipeline_test.cc b/dali/pipeline/pipeline_test.cc index 0f4261c4689..705c455acde 100644 --- a/dali/pipeline/pipeline_test.cc +++ b/dali/pipeline/pipeline_test.cc @@ -307,9 +307,9 @@ class DummyPresizeOpCPU : public Operator { } void RunImpl(HostWorkspace &ws) override { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); int num_samples = input.shape().num_samples(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); auto tmp_size = output.capacity(); output.set_type(); output.Resize(uniform_list_shape(num_samples, std::vector{2})); @@ -332,9 +332,9 @@ class DummyPresizeOpGPU : public Operator { } void RunImpl(DeviceWorkspace &ws) override { - const auto &input = ws.InputRef(0); + const auto &input = ws.Input(0); int num_samples = input.shape().num_samples(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.set_type(); size_t tmp_size[2] = {output.capacity(), input.capacity()}; output.Resize(uniform_list_shape(num_samples, std::vector{2})); @@ -358,9 +358,9 @@ class DummyPresizeOpMixed : public Operator { using Operator::Run; void Run(MixedWorkspace &ws) override { - auto &input = ws.InputRef(0); + auto &input = ws.Input(0); int num_samples = input.shape().num_samples(); - auto &output = ws.OutputRef(0); + auto &output = ws.Output(0); output.set_type(); size_t tmp_size[2] = {output.capacity(), input.capacity()}; output.Resize(uniform_list_shape(num_samples, std::vector{2})); @@ -465,29 +465,29 @@ TEST_F(PipelineTestOnce, TestPresize) { pipe.Outputs(&ws); // we should not presize CPU buffers if they are not pinned - ASSERT_EQ(*(ws.OutputRef(0).tensor(0)), 0); + ASSERT_EQ(*(ws.Output(0).tensor(0)), 0); int ref_presize = RestrictPinnedMemUsage() ? 0 : presize_val_CPU; - ASSERT_EQ(*(ws.OutputRef(1).tensor(0)), ref_presize); + ASSERT_EQ(*(ws.Output(1).tensor(0)), ref_presize); size_t tmp[2]; CUDA_CALL(cudaDeviceSynchronize()); - CUDA_CALL(cudaMemcpy(&tmp, ws.OutputRef(2).tensor(0), + CUDA_CALL(cudaMemcpy(&tmp, ws.Output(2).tensor(0), sizeof(size_t) * 2, cudaMemcpyDefault)); ASSERT_EQ(tmp[0], presize_val_Mixed); ASSERT_EQ(tmp[1], 2 * sizeof(size_t)); - CUDA_CALL(cudaMemcpy(&tmp, ws.OutputRef(3).tensor(0), + CUDA_CALL(cudaMemcpy(&tmp, ws.Output(3).tensor(0), sizeof(size_t) * 2, cudaMemcpyDefault)); ASSERT_EQ(tmp[0], presize_val_GPU); ASSERT_EQ(tmp[1], 2 * sizeof(size_t)); - CUDA_CALL(cudaMemcpy(&tmp, ws.OutputRef(4).tensor(0), + CUDA_CALL(cudaMemcpy(&tmp, ws.Output(4).tensor(0), sizeof(size_t) * 2, cudaMemcpyDefault)); ASSERT_EQ(tmp[0], presize_val_GPU); ASSERT_EQ(tmp[1], 2 * sizeof(size_t)); - CUDA_CALL(cudaMemcpy(&tmp, ws.OutputRef(5).tensor(0), + CUDA_CALL(cudaMemcpy(&tmp, ws.Output(5).tensor(0), sizeof(size_t) * 2, cudaMemcpyDefault)); ASSERT_EQ(tmp[0], presize_val_default); ASSERT_EQ(tmp[1], 2 * sizeof(size_t)); @@ -551,7 +551,7 @@ class PrefetchedPipelineTest : public GenericDecoderTest { ASSERT_EQ(ws.NumOutput(), 1); ASSERT_EQ(ws.NumInput(), 0); ASSERT_TRUE(ws.OutputIsType(0)); - TensorList &res1 = ws.OutputRef(0); + TensorList &res1 = ws.Output(0); for (int j = 0; j < batch_size; ++j) { this->VerifyDecode( res1.template tensor(j), diff --git a/dali/pipeline/workspace/sample_workspace.cc b/dali/pipeline/workspace/sample_workspace.cc index 7f9bcf42a56..eaae1ba1170 100644 --- a/dali/pipeline/workspace/sample_workspace.cc +++ b/dali/pipeline/workspace/sample_workspace.cc @@ -24,10 +24,10 @@ void MakeSampleView(SampleWorkspace& sample, HostWorkspace& batch, int data_idx, int num_inputs = batch.NumInput(); for (int i = 0; i < num_inputs; i++) { if (batch.InputIsType(i)) { - auto &input_ref = batch.InputRef(i); + auto &input_ref = batch.Input(i); sample.AddInput(&input_ref[data_idx]); } else { - auto &input_ref = batch.InputRef(i); + auto &input_ref = batch.Input(i); sample.AddInput(&input_ref[data_idx]); } } @@ -35,10 +35,10 @@ void MakeSampleView(SampleWorkspace& sample, HostWorkspace& batch, int data_idx, int num_outputs = batch.NumOutput(); for (int i = 0; i < num_outputs; i++) { if (batch.OutputIsType(i)) { - auto &output_ref = batch.OutputRef(i); + auto &output_ref = batch.Output(i); sample.AddOutput(&output_ref[data_idx]); } else { - auto &output_ref = batch.OutputRef(i); + auto &output_ref = batch.Output(i); sample.AddOutput(&output_ref[data_idx]); } } diff --git a/dali/pipeline/workspace/workspace.h b/dali/pipeline/workspace/workspace.h index 6d53ad6ae92..394eebc1e10 100644 --- a/dali/pipeline/workspace/workspace.h +++ b/dali/pipeline/workspace/workspace.h @@ -142,12 +142,12 @@ class WorkspaceBase : public ArgumentWorkspace { } template - auto& InputRef(int idx) const { + auto& Input(int idx) const { return *InputHandle(idx, Backend{}); } template - auto& OutputRef(int idx) const { + auto& Output(int idx) const { return *OutputHandle(idx, Backend{}); } @@ -181,9 +181,9 @@ class WorkspaceBase : public ArgumentWorkspace { */ auto GetInputShape(int input_idx) const { if (InputIsType(input_idx)) { - return InputRef(input_idx).shape(); + return Input(input_idx).shape(); } else { - return InputRef(input_idx).shape(); + return Input(input_idx).shape(); } } @@ -195,9 +195,9 @@ class WorkspaceBase : public ArgumentWorkspace { DALI_ENFORCE(input_idx >= 0 && input_idx < NumInput(), make_string("Invalid input index: ", input_idx, "; while NumInput: ", NumInput())); if (InputIsType(input_idx)) { - return InputRef(input_idx).num_samples(); + return Input(input_idx).num_samples(); } else { - return InputRef(input_idx).num_samples(); + return Input(input_idx).num_samples(); } } @@ -209,9 +209,9 @@ class WorkspaceBase : public ArgumentWorkspace { DALI_ENFORCE(input_idx >= 0 && input_idx < NumInput(), make_string("Invalid input index: ", input_idx, "; while NumInput: ", NumInput())); if (InputIsType(input_idx)) { - return InputRef(input_idx).sample_dim(); + return Input(input_idx).sample_dim(); } else { - return InputRef(input_idx).sample_dim(); + return Input(input_idx).sample_dim(); } } diff --git a/dali/test/dali_operator_test.h b/dali/test/dali_operator_test.h index 15a11cc9e36..693eb21f51c 100644 --- a/dali/test/dali_operator_test.h +++ b/dali/test/dali_operator_test.h @@ -142,9 +142,9 @@ GetOutputsFromPipeline(Pipeline &pipeline, const std::string &output_backend) { pipeline.Outputs(&workspace); for (int output_idx = 0; output_idx < workspace.NumOutput(); output_idx++) { if (workspace.OutputIsType(output_idx)) { - ret.emplace_back(&workspace.template OutputRef(output_idx)); + ret.emplace_back(&workspace.template Output(output_idx)); } else { - ret.emplace_back(&workspace.template OutputRef(output_idx)); + ret.emplace_back(&workspace.template Output(output_idx)); } } return ret; diff --git a/dali/test/dali_test_bboxes.h b/dali/test/dali_test_bboxes.h index e7956471abe..1ea10275d53 100644 --- a/dali/test/dali_test_bboxes.h +++ b/dali/test/dali_test_bboxes.h @@ -95,11 +95,11 @@ class GenericBBoxesTest : public DALISingleOpTest { DeviceWorkspace ws; pipe->Outputs(&ws); - auto images_cpu = this->CopyToHost(ws.OutputRef(0))[0]; - images_cpu->SetLayout(ws.OutputRef(0).GetLayout()); + auto images_cpu = this->CopyToHost(ws.Output(0))[0]; + images_cpu->SetLayout(ws.Output(0).GetLayout()); - auto boxes_cpu = this->CopyToHost(ws.OutputRef(1))[0]; - boxes_cpu->SetLayout(ws.OutputRef(1).GetLayout()); + auto boxes_cpu = this->CopyToHost(ws.Output(1))[0]; + boxes_cpu->SetLayout(ws.Output(1).GetLayout()); return {images_cpu, boxes_cpu}; } @@ -145,15 +145,15 @@ class GenericBBoxesTest : public DALISingleOpTest { std::vector>> ret; ret.push_back(std::make_shared>()); ret.push_back(std::make_shared>()); - ret[0]->Copy(ws.OutputRef(0), 0); - ret[1]->Copy(ws.OutputRef(1), 0); + ret[0]->Copy(ws.Output(0), 0); + ret[1]->Copy(ws.Output(1), 0); return ret; } vector>> Reference( const vector *> &inputs, DeviceWorkspace *ws) override { - auto &from = ws->OutputRef(1); + auto &from = ws->Output(1); auto reference = this->CopyToHost(from); reference[0]->SetLayout(from.GetLayout()); return reference; diff --git a/dali/test/dali_test_matching.h b/dali/test/dali_test_matching.h index 134ee8cf274..a97ac7b647d 100644 --- a/dali/test/dali_test_matching.h +++ b/dali/test/dali_test_matching.h @@ -59,9 +59,9 @@ class GenericMatchingTest : public DALISingleOpTest { vector>> Reference(const vector*> &inputs, DeviceWorkspace *ws) override { if (GetOpType() == OpType::GPU) - return this->CopyToHost(ws->OutputRef(1)); + return this->CopyToHost(ws->Output(1)); else - return this->CopyToHost(ws->OutputRef(1)); + return this->CopyToHost(ws->Output(1)); } uint32_t GetTestCheckType() const override { diff --git a/dali/test/dali_test_resize.h b/dali/test/dali_test_resize.h index d31b944e398..ce74789bfc5 100755 --- a/dali/test/dali_test_resize.h +++ b/dali/test/dali_test_resize.h @@ -1,4 +1,4 @@ -// Copyright (c) 2017-2018, NVIDIA CORPORATION. All rights reserved. +// Copyright (c) 2017-2021, NVIDIA CORPORATION & AFFILIATES. All rights reserved. #ifndef DALI_TEST_DALI_TEST_RESIZE_H_ #define DALI_TEST_DALI_TEST_RESIZE_H_ @@ -76,7 +76,7 @@ class GenericResizeTest : public DALISingleOpTest { // determine resize parameters if (useExternSizes) { - const auto *t = ws->OutputRef(1).tensor(i); + const auto *t = ws->Output(1).tensor(i); rsz_h = t[0]; rsz_w = t[1]; } else { diff --git a/dali/test/dali_test_single_op.h b/dali/test/dali_test_single_op.h index 9e2c2e464ae..cc3d484d229 100644 --- a/dali/test/dali_test_single_op.h +++ b/dali/test/dali_test_single_op.h @@ -308,9 +308,9 @@ class DALISingleOpTest : public DALITest { auto idx = output_indices[i]; if (output_device == "gpu") { // copy to host - calc_output->Copy(ws->OutputRef(idx), nullptr); + calc_output->Copy(ws->Output(idx), nullptr); } else { - calc_output->Copy(ws->OutputRef(idx), nullptr); + calc_output->Copy(ws->Output(idx), nullptr); } auto& ref_output = res[i]; diff --git a/dali/test/plugins/dummy/dummy.cc b/dali/test/plugins/dummy/dummy.cc index 539dadab4c2..dd018bdf493 100644 --- a/dali/test/plugins/dummy/dummy.cc +++ b/dali/test/plugins/dummy/dummy.cc @@ -18,8 +18,8 @@ namespace other_ns { template <> void Dummy<::dali::CPUBackend>::RunImpl(::dali::HostWorkspace &ws) { - const auto &input = ws.InputRef<::dali::CPUBackend>(0); - auto &output = ws.OutputRef<::dali::CPUBackend>(0); + const auto &input = ws.Input<::dali::CPUBackend>(0); + auto &output = ws.Output<::dali::CPUBackend>(0); ::dali::TypeInfo type = input.type_info(); auto &tp = ws.GetThreadPool(); diff --git a/dali/test/plugins/dummy/dummy.cu b/dali/test/plugins/dummy/dummy.cu index a11e6a51a58..10409ead010 100644 --- a/dali/test/plugins/dummy/dummy.cu +++ b/dali/test/plugins/dummy/dummy.cu @@ -19,9 +19,9 @@ namespace other_ns { template<> void Dummy<::dali::GPUBackend>::RunImpl(::dali::DeviceWorkspace &ws) { - const auto &input = ws.InputRef<::dali::GPUBackend>(0); + const auto &input = ws.Input<::dali::GPUBackend>(0); const auto &shape = input.shape(); - auto &output = ws.OutputRef<::dali::GPUBackend>(0); + auto &output = ws.Output<::dali::GPUBackend>(0); for (int sample_idx = 0; sample_idx < shape.num_samples(); sample_idx++) { CUDA_CALL(cudaMemcpyAsync( output.raw_mutable_tensor(sample_idx), diff --git a/dali/test/plugins/dummy/dummy.h b/dali/test/plugins/dummy/dummy.h index 780b081e1a3..9fa4e5752aa 100644 --- a/dali/test/plugins/dummy/dummy.h +++ b/dali/test/plugins/dummy/dummy.h @@ -39,7 +39,7 @@ class Dummy : public ::dali::Operator { bool SetupImpl(std::vector<::dali::OutputDesc> &output_desc, const ::dali::workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc.resize(1); output_desc[0] = {input.shape(), input.type()}; return true; diff --git a/docs/examples/custom_operations/custom_operator/create_a_custom_operator.ipynb b/docs/examples/custom_operations/custom_operator/create_a_custom_operator.ipynb index 4e2385c3105..6fface74b5b 100644 --- a/docs/examples/custom_operations/custom_operator/create_a_custom_operator.ipynb +++ b/docs/examples/custom_operations/custom_operator/create_a_custom_operator.ipynb @@ -72,7 +72,7 @@ "\r\n", " bool SetupImpl(std::vector<::dali::OutputDesc> &output_desc,\r\n", " const ::dali::workspace_t &ws) override {\r\n", - " const auto &input = ws.template InputRef(0);\r\n", + " const auto &input = ws.template Input(0);\r\n", " output_desc.resize(1);\r\n", " output_desc[0] = {input.shape(), input.type()};\r\n", " return true;\r\n", @@ -121,8 +121,8 @@ "\r\n", "template <>\r\n", "void Dummy<::dali::CPUBackend>::RunImpl(::dali::HostWorkspace &ws) {\r\n", - " const auto &input = ws.InputRef<::dali::CPUBackend>(0);\r\n", - " auto &output = ws.OutputRef<::dali::CPUBackend>(0);\r\n", + " const auto &input = ws.Input<::dali::CPUBackend>(0);\r\n", + " auto &output = ws.Output<::dali::CPUBackend>(0);\r\n", "\r\n", " ::dali::TypeInfo type = input.type_info();\r\n", " auto &tp = ws.GetThreadPool();\r\n", @@ -185,9 +185,9 @@ "\r\n", "template<>\r\n", "void Dummy<::dali::GPUBackend>::RunImpl(::dali::DeviceWorkspace &ws) {\r\n", - " const auto &input = ws.InputRef<::dali::GPUBackend>(0);\r\n", + " const auto &input = ws.Input<::dali::GPUBackend>(0);\r\n", " const auto &shape = input.shape();\r\n", - " auto &output = ws.OutputRef<::dali::GPUBackend>(0);\r\n", + " auto &output = ws.Output<::dali::GPUBackend>(0);\r\n", " for (int sample_idx = 0; sample_idx < shape.num_samples(); sample_idx++) {\r\n", " CUDA_CALL(cudaMemcpyAsync(\r\n", " output.raw_mutable_tensor(sample_idx),\r\n", diff --git a/docs/examples/custom_operations/custom_operator/customdummy/dummy.cc b/docs/examples/custom_operations/custom_operator/customdummy/dummy.cc index f6e3bc3771e..3a8e00e5098 100644 --- a/docs/examples/custom_operations/custom_operator/customdummy/dummy.cc +++ b/docs/examples/custom_operations/custom_operator/customdummy/dummy.cc @@ -4,8 +4,8 @@ namespace other_ns { template <> void Dummy<::dali::CPUBackend>::RunImpl(::dali::HostWorkspace &ws) { - const auto &input = ws.InputRef<::dali::CPUBackend>(0); - auto &output = ws.OutputRef<::dali::CPUBackend>(0); + const auto &input = ws.Input<::dali::CPUBackend>(0); + auto &output = ws.Output<::dali::CPUBackend>(0); ::dali::TypeInfo type = input.type_info(); auto &tp = ws.GetThreadPool(); diff --git a/docs/examples/custom_operations/custom_operator/customdummy/dummy.cu b/docs/examples/custom_operations/custom_operator/customdummy/dummy.cu index 1b8a01958e9..ab3fad57d18 100644 --- a/docs/examples/custom_operations/custom_operator/customdummy/dummy.cu +++ b/docs/examples/custom_operations/custom_operator/customdummy/dummy.cu @@ -5,9 +5,9 @@ namespace other_ns { template<> void Dummy<::dali::GPUBackend>::RunImpl(::dali::DeviceWorkspace &ws) { - const auto &input = ws.InputRef<::dali::GPUBackend>(0); + const auto &input = ws.Input<::dali::GPUBackend>(0); const auto &shape = input.shape(); - auto &output = ws.OutputRef<::dali::GPUBackend>(0); + auto &output = ws.Output<::dali::GPUBackend>(0); for (int sample_idx = 0; sample_idx < shape.num_samples(); sample_idx++) { CUDA_CALL(cudaMemcpyAsync( output.raw_mutable_tensor(sample_idx), diff --git a/docs/examples/custom_operations/custom_operator/customdummy/dummy.h b/docs/examples/custom_operations/custom_operator/customdummy/dummy.h index 8d7da2d7df3..a4f29dd7bff 100644 --- a/docs/examples/custom_operations/custom_operator/customdummy/dummy.h +++ b/docs/examples/custom_operations/custom_operator/customdummy/dummy.h @@ -27,7 +27,7 @@ class Dummy : public ::dali::Operator { bool SetupImpl(std::vector<::dali::OutputDesc> &output_desc, const ::dali::workspace_t &ws) override { - const auto &input = ws.template InputRef(0); + const auto &input = ws.template Input(0); output_desc.resize(1); output_desc[0] = {input.shape(), input.type()}; return true;