m53297601
/
mindspore
1
0
Fork 0
Code Issues Pull Requests Releases Wiki Activity

!1254 change runtime stream type from uintptr_t to void*

Browse Source 
Merge pull request !1254 from caifubi/change-uintprt_t-to-void-ptr
pull/1254/MERGE
mindspore-ci-bot 5 years ago committed by Gitee
parent ab6bc60d98 5b963aef2b
commit c964916e20
111 changed files with 142 additions and 155 deletions
Show Stats Download Patch File Download Diff File

2
mindspore/ccsrc/device/gpu/gpu_kernel_runtime.cc
Unescape View File

@ -199,7 +199,7 @@ bool GPUKernelRuntime::LaunchKernelDynamic(const session::KernelGraph *graph) {
AddressPtrList kernel_workspaces;
AddressPtrList kernel_outputs;
AllocKernelDynamicRes(*kernel_mod, kernel, &kernel_inputs, &kernel_workspaces, &kernel_outputs);
if (!kernel_mod->Launch(kernel_inputs, kernel_workspaces, kernel_outputs, reinterpret_cast<uintptr_t>(stream_))) {
if (!kernel_mod->Launch(kernel_inputs, kernel_workspaces, kernel_outputs, stream_)) {
MS_LOG(ERROR) << "Launch kernel failed.";
return false;
}

3
mindspore/ccsrc/device/kernel_runtime.cc
Unescape View File

@ -664,8 +664,7 @@ bool KernelRuntime::LaunchKernelMod(const session::KernelGraph &graph) {
struct timeval start_time, end_time;
(void)gettimeofday(&start_time, nullptr);
#endif
auto ret =
kernel_mod->Launch(kernel_inputs, kernel_workspaces, kernel_outputs, reinterpret_cast<uintptr_t>(stream_));
auto ret = kernel_mod->Launch(kernel_inputs, kernel_workspaces, kernel_outputs, stream_);
if (!ret) {
MS_LOG(ERROR) << "Launch kernel failed.";
return false;

7
mindspore/ccsrc/kernel/aicpu/aicpu_kernel_mod.cc
Unescape View File

@ -103,14 +103,13 @@ void AicpuOpKernelMod::CreateCpuKernelInfo(const std::vector<AddressPtr> &inputs
}
bool AicpuOpKernelMod::Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) {
if (stream_ptr == 0) {
const std::vector<AddressPtr> &outputs, void *stream_ptr) {
if (stream_ptr == nullptr) {
MS_LOG(ERROR) << "stream_ptr should not be nullptr.";
return false;
}
CreateCpuKernelInfo(inputs, outputs);
auto *stream = reinterpret_cast<rtStream_t *>(stream_ptr);
if (node_name_ == "TopK") {
node_name_ = "TopKV2";
}
@ -119,7 +118,7 @@ bool AicpuOpKernelMod::Launch(const std::vector<AddressPtr> &inputs, const std::
if (rtCpuKernelLaunch(reinterpret_cast<const void *>(node_so_.c_str()),
reinterpret_cast<const void *>(node_name_.c_str()), 1,
reinterpret_cast<const void *>(args_.data()), static_cast<uint32_t>(args_.length()), nullptr,
stream) != RT_ERROR_NONE) {
stream_ptr) != RT_ERROR_NONE) {
MS_LOG(ERROR) << "Aicpu op launch failed!";
return false;

2
mindspore/ccsrc/kernel/aicpu/aicpu_kernel_mod.h
Unescape View File

@ -27,7 +27,7 @@ class AicpuOpKernelMod : public AscendKernelMod {
AicpuOpKernelMod();
~AicpuOpKernelMod() override;
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override;
const std::vector<AddressPtr> &outputs, void *stream_ptr) override;
std::vector<TaskInfoPtr> GenTask(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uint32_t stream_id) override;

2
mindspore/ccsrc/kernel/akg/gpu/akg_gpu_kernel_mod.cc
Unescape View File

@ -82,7 +82,7 @@ const std::vector<size_t> &GpuKernelMod::GetOutputSizeList() const { return outp
const std::vector<size_t> &GpuKernelMod::GetWorkspaceSizeList() const { return workspace_size_list_; }
bool GpuKernelMod::Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) {
const std::vector<AddressPtr> &outputs, void *stream_ptr) {
if (stream_ptr == 0) {
MS_LOG(ERROR) << "stream_ptr should not be nullptr.";
return false;

2
mindspore/ccsrc/kernel/akg/gpu/akg_gpu_kernel_mod.h
Unescape View File

@ -64,7 +64,7 @@ class GpuKernelMod : public KernelMod {
const std::vector<size_t> &GetOutputSizeList() const override;
const std::vector<size_t> &GetWorkspaceSizeList() const override;
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override;
const std::vector<AddressPtr> &outputs, void *stream_ptr) override;
static GpuKernelManagerPtr kernelmanager_;

2
mindspore/ccsrc/kernel/cpu/cpu_kernel.h
Unescape View File

@ -56,7 +56,7 @@ class CPUKernel : public kernel::KernelMod {
void Init(const CNodePtr &kernel_node);
virtual void InitKernel(const CNodePtr &kernel_node) = 0;
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t /*stream_ptr*/) override {
const std::vector<AddressPtr> &outputs, void * /*stream_ptr*/) override {
return Launch(inputs, workspace, outputs);
};
virtual bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,

2
mindspore/ccsrc/kernel/gpu/arrays/argmax_gpu_kernel.h
Unescape View File

@ -35,7 +35,7 @@ class ArgmaxGpuKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
T *input = GetDeviceAddress<T>(inputs, 0);
int *output = GetDeviceAddress<int>(outputs, 0);
CalArgmax(input, SizeToInt(batch_size_), SizeToInt(channel_size_), axis_, output,

2
mindspore/ccsrc/kernel/gpu/arrays/array_reduce_gpu_kernel.h
Unescape View File

@ -55,7 +55,7 @@ class ArrayReduceGpuKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
if (is_null_input_) {
return true;
}

2
mindspore/ccsrc/kernel/gpu/arrays/concatv2_gpu_kernel.h
Unescape View File

@ -34,7 +34,7 @@ class ConcatV2GpuFwdKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
if (inputs.size() == 2) {
T *input_0 = GetDeviceAddress<T>(inputs, 0);
T *input_1 = GetDeviceAddress<T>(inputs, 1);

2
mindspore/ccsrc/kernel/gpu/arrays/gather_gpu_kernel.h
Unescape View File

@ -35,7 +35,7 @@ class GatherGpuFwdKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
VARIABLE_NOT_USED(workspace);
T *input_addr = GetDeviceAddress<T>(inputs, 0);
S *indices_addr = GetDeviceAddress<S>(inputs, 1);

2
mindspore/ccsrc/kernel/gpu/arrays/one_hot_gpu_kernel.h
Unescape View File

@ -34,7 +34,7 @@ class OneHotGpuFwdKernel : public GpuKernel {
const std::vector<size_t> &GetOutputSizeList() const override { return output_size_list_; }
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
VARIABLE_NOT_USED(workspace);
const S *indices = GetDeviceAddress<S>(inputs, 0);
const T *on_value = GetDeviceAddress<T>(inputs, 1);

2
mindspore/ccsrc/kernel/gpu/arrays/select_gpu_kernel.h
Unescape View File

@ -34,7 +34,7 @@ class SelectGpuKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
bool *input_cond = GetDeviceAddress<bool>(inputs, 0);
T *input_x = GetDeviceAddress<T>(inputs, 1);
T *input_y = GetDeviceAddress<T>(inputs, 2);

2
mindspore/ccsrc/kernel/gpu/arrays/slice_gpu_kernel.h
Unescape View File

@ -34,7 +34,7 @@ class SliceGpuFwdKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
T *input = GetDeviceAddress<T>(inputs, 0);
T *output = GetDeviceAddress<T>(outputs, 0);
if (is_strided_slice_) {

2
mindspore/ccsrc/kernel/gpu/arrays/slice_grad_gpu_kernel.h
Unescape View File

@ -34,7 +34,7 @@ class SliceGradGpuKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
T *dy = GetDeviceAddress<T>(inputs, 0);
T *dx = GetDeviceAddress<T>(outputs, 0);
FillDeviceArray(outputs[0]->size / sizeof(T), dx, 0.f, reinterpret_cast<cudaStream_t>(stream_ptr));

2
mindspore/ccsrc/kernel/gpu/arrays/transpose_gpu_kernel.h
Unescape View File

@ -34,7 +34,7 @@ class TransposeGpuFwdKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
T *input = GetDeviceAddress<T>(inputs, 0);
T *output = GetDeviceAddress<T>(outputs, 0);
int *input_shape = GetDeviceAddress<int>(workspace, 0);

2
mindspore/ccsrc/kernel/gpu/arrays/unsorted_segment_sum_gpu_kernel.h
Unescape View File

@ -35,7 +35,7 @@ class UnsortedSegmentSumGpuKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
T *input_addr = GetDeviceAddress<T>(inputs, 0);
S *indices_addr = GetDeviceAddress<S>(inputs, 1);
T *output_addr = GetDeviceAddress<T>(outputs, 0);

2
mindspore/ccsrc/kernel/gpu/control/recv_gpu_kernel.h
Unescape View File

@ -33,7 +33,7 @@ class RecvGpuKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &, const std::vector<AddressPtr> &, const std::vector<AddressPtr> &,
uintptr_t) override {
void *) override {
CHECK_CUDA_RET_WITH_EXCEPT(cudaStreamWaitEvent(wait_stream_, wait_event_, 0), "Waiting cuda event failed.");
return true;
}

2
mindspore/ccsrc/kernel/gpu/control/send_gpu_kernel.h
Unescape View File

@ -33,7 +33,7 @@ class SendGpuKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &, const std::vector<AddressPtr> &, const std::vector<AddressPtr> &,
uintptr_t) override {
void *) override {
CHECK_CUDA_RET_WITH_EXCEPT(cudaEventRecord(record_event_, record_stream_), "Recording cuda event failed.");
return true;
}

2
mindspore/ccsrc/kernel/gpu/data/dataset_init_kernel.cc
Unescape View File

@ -53,7 +53,7 @@ bool DatasetInitKernel::Init(const CNodePtr &kernel_node) {
void DatasetInitKernel::InitSizeLists() { return; }
bool DatasetInitKernel::Launch(const std::vector<AddressPtr> &, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &, uintptr_t) {
const std::vector<AddressPtr> &, void *) {
void *addr = nullptr;
size_t len = total_bytes_ * buffer_q_capacity_;

2
mindspore/ccsrc/kernel/gpu/data/dataset_init_kernel.h
Unescape View File

@ -33,7 +33,7 @@ class DatasetInitKernel : public GpuKernel {
const std::vector<size_t> &GetOutputSizeList() const override;
const std::vector<size_t> &GetWorkspaceSizeList() const override;
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override;
const std::vector<AddressPtr> &outputs, void *stream_ptr) override;
bool Init(const CNodePtr &kernel_node) override;
protected:

2
mindspore/ccsrc/kernel/gpu/data/dataset_iterator_kernel.cc
Unescape View File

@ -64,7 +64,7 @@ bool DatasetIteratorKernel::Init(const CNodePtr &kernel_node) {
void DatasetIteratorKernel::InitSizeLists() { return; }
bool DatasetIteratorKernel::Launch(const std::vector<AddressPtr> &, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t) {
const std::vector<AddressPtr> &outputs, void *) {
void *addr = nullptr;
size_t len = 0;

2
mindspore/ccsrc/kernel/gpu/data/dataset_iterator_kernel.h
Unescape View File

@ -33,7 +33,7 @@ class DatasetIteratorKernel : public GpuKernel {
const std::vector<size_t> &GetOutputSizeList() const override;
const std::vector<size_t> &GetWorkspaceSizeList() const override;
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &workspace,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override;
const std::vector<AddressPtr> &outputs, void *stream_ptr) override;
bool Init(const CNodePtr &kernel_node) override;
protected:

2
mindspore/ccsrc/kernel/gpu/math/addn_gpu_kernel.h
Unescape View File

@ -43,7 +43,7 @@ class AddNGpuFwdKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t) override {
const std::vector<AddressPtr> &outputs, void *) override {
if (is_null_input_) {
return true;
}

2
mindspore/ccsrc/kernel/gpu/math/assign_add_gpu_kernel.h
Unescape View File

@ -35,7 +35,7 @@ class AssignAddGpuFwdKernel : public GpuKernel {
const std::vector<size_t> &GetWorkspaceSizeList() const override { return workspace_size_list_; }
bool Launch(const std::vector<AddressPtr> &inputs, const std::vector<AddressPtr> &,
const std::vector<AddressPtr> &outputs, uintptr_t stream_ptr) override {
const std::vector<AddressPtr> &outputs, void *stream_ptr) override {
if (is_null_input_) {
return true;
}

Some files were not shown because too many files have changed in this diff Show More

Write Preview
Loading…
Cancel
Save