Skip to content
New issue

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

Sign up for GitHub

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

Already on GitHub? Sign in to your account

Jump to bottom

[dual expand dim]Init dual expand dim test #23

Open
wants to merge 4 commits into
base: paddlebox
Choose a base branch
from
Open

[dual expand dim]Init dual expand dim test #23

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
4 commits
Select commit Hold shift + click to select a range
6f18317
init for dual expand dim
zmxdream Nov 4, 2023
9bc7fd0
init for dual expand dim
zmxdream Nov 4, 2023
1e0e1cc
Merge remote-tracking branch 'origin/paddlebox' into init_dual_expand…
zmxdream Nov 4, 2023
f50a895
init for dual expand dim
zmxdream Nov 4, 2023
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
12 changes: 12 additions & 0 deletions paddle/fluid/framework/data_feed.cc
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -3609,6 +3609,18 @@ void SlotPaddleBoxDataFeed::GetRankOffsetGPU(const int pv_num,
value.d_ad_offset.data<int>(), col);

#elif defined(PADDLE_WITH_XPU_KP)
auto dev_ctx = platform::DeviceContextPool::Instance().Get(this->place_);
auto ctx = static_cast<platform::XPUDeviceContext*>(dev_ctx)->x_context();
int r = xpu::constant<int>(ctx, tensor_ptr, rank_offset_->numel(), 0);
PADDLE_ENFORCE_EQ(r,
XPU_SUCCESS,
platform::errors::External(
"XPU constant kernel return wrong value[%d %s]",
r,
XPUAPIErrorMsg[r]));
// DataFeedPdboxXpuKernelHelper::CopyRankOffset(this->place_, tensor_ptr, ins_num, pv_num, max_rank,
// value.d_rank.data<int>(), value.d_cmatch.data<int>(),
// value.d_ad_offset.data<int>(), col);
DataFeedPdboxXpuKernelHelper::CopyRankOffset(this->place_, tensor_ptr, ins_num, pv_num, max_rank,
value.d_rank.data<int>(), value.d_cmatch.data<int>(),
value.d_ad_offset.data<int>(), col);
Expand Down
5 changes: 3 additions & 2 deletions paddle/fluid/framework/data_feed.kps
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -376,7 +376,8 @@ void DataFeedPdboxXpuKernelHelper::CopyRankOffset(const paddle::platform::Place&
stream = static_cast<platform::XPUDeviceContext*>(dev_ctx)
->x_context()
->xpu_stream;
CopyRankOffsetKernel<<<4, 64, stream>>>(dest, ranks, cmatchs, ad_offsets, ins_num, pv_num, max_rank, cols);
// CopyRankOffsetKernel<<<4, 64, stream>>>(dest, ranks, cmatchs, ad_offsets, ins_num, pv_num, max_rank, cols);
CopyRankOffsetKernel<<<8, 64, stream>>>(dest, ranks, cmatchs, ad_offsets, pv_num, ins_num, max_rank, cols);
xpu_wait(stream);
}

Expand Down Expand Up @@ -430,4 +431,4 @@ void DataFeedPdboxXpuKernelHelper::CopyForTensor(const paddle::platform::Place&

} // end namespace framework
} // end namespace paddle
#endif
#endif
18 changes: 9 additions & 9 deletions paddle/fluid/framework/fleet/box_wrapper_impl.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -731,16 +731,16 @@ void BoxWrapper::PushSparseGradCaseXPU(const paddle::platform::Place& place,

TRACE_SCOPE_START("CopyForPush", xpu_wait(ctx_xpu->xpu_stream));
#endif
float* real_grad_values;
for (int i = 0; i < slot_num; i++) {
if(grad_values[i] != nullptr) {
real_grad_values = const_cast<float*>(grad_values[i]);
break;
}
}
box_wrapper_kernel_->CopyForPush(place, real_grad_values, total_grad_values_xpu,
// float* real_grad_values;
// for (int i = 0; i < slot_num; i++) {
// if(grad_values[i] != nullptr) {
// real_grad_values = const_cast<float*>(grad_values[i]);
// break;
// }
// }
box_wrapper_kernel_->CopyForPush(place, xpu_values, total_grad_values_xpu,
push_offset, total_length, slot_vector, slot_lens, slot_num,
hidden_size, batch_size, total_dims, skip_offset, key2slot);
hidden_size, expand_embed_dim, batch_size, total_dims, skip_offset, key2slot, expand_only);

push_boxps_timer.Resume();
#ifdef TRACE_PROFILE
Expand Down
6 changes: 4 additions & 2 deletions paddle/fluid/framework/fleet/box_wrapper_kernel.h
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -48,18 +48,20 @@ void CopyForPull(

void CopyForPush(
const paddle::platform::Place& place,
float* gm_src_ptr,
float** gm_src_ptr,
void* total_grad_values_xpu,
boxps::FeaturePushOffset* push_offset,
const int64_t total_length,
const int* slots,
const int64_t* slot_lens,
const int slot_num,
const int hidden_size,
const int expand_embed_dim,
const int batch_size,
const int* total_dims,
const int skip_offset,
const int* key2slot);
const int* key2slot,
bool expand_only);

public:
const static int MAX_SLOT_SIZE = 10240;
Expand Down
Loading