From b077291f03db4c304de4a8814ba5cf9fc5007f52 Mon Sep 17 00:00:00 2001 From: "haowen.han@mthreads.com" Date: Mon, 13 May 2024 13:29:24 +0800 Subject: [PATCH] Revert "fix the unqiue op that generate the wrong the inreverse result (#62104)" This reverts commit b89066ae29ab20f9c35e10f52093613d505d2f83. --- paddle/phi/kernels/gpu/unique_kernel.cu | 11 +---------- 1 file changed, 1 insertion(+), 10 deletions(-) diff --git a/paddle/phi/kernels/gpu/unique_kernel.cu b/paddle/phi/kernels/gpu/unique_kernel.cu index 093876a402763..59501150d6479 100644 --- a/paddle/phi/kernels/gpu/unique_kernel.cu +++ b/paddle/phi/kernels/gpu/unique_kernel.cu @@ -168,11 +168,8 @@ UniqueFlattendCUDATensor(const Context& context, #ifdef PADDLE_WITH_HIP hipMemset(inv_loc_data_ptr, 0, sizeof(IndexT)); #else - thrust::device_ptr inv_loc_data_dev(inv_loc_data_ptr); - inv_loc_data_dev[0] = 0; // without device_ptr, segmentation fault + cudaMemsetAsync(inv_loc_data_ptr, 0, sizeof(IndexT), context.stream()); #endif - -#ifdef PADDLE_WITH_HIP size_t temp_storage_bytes = 0; cub::DeviceScan::InclusiveSum(NULL, temp_storage_bytes, @@ -188,12 +185,6 @@ UniqueFlattendCUDATensor(const Context& context, inv_loc_data_ptr, num_input, context.stream()); -#else - thrust::inclusive_scan(exec_policy, - inv_loc_data_ptr, - inv_loc_data_ptr + num_input, - inv_loc_data_ptr); -#endif thrust::scatter(exec_policy, inv_loc_data_ptr, inv_loc_data_ptr + num_input,