| 3506 | } |
| 3507 | |
| 3508 | bool MIOpenSupport::DoBiasAdd(Stream* stream, |
| 3509 | const DeviceMemory<float>& input_data, |
| 3510 | const DeviceMemory<float>& biases, |
| 3511 | const dnn::BatchDescriptor& dimensions, |
| 3512 | DeviceMemory<float>* output_data) { |
| 3513 | ScopedTensorDescriptor input_descriptor{dimensions, miopenFloat}; |
| 3514 | |
| 3515 | BatchDescriptor bias_dimensions; |
| 3516 | bias_dimensions.set_count(1) |
| 3517 | .set_feature_map_count(dimensions.feature_map_count()) |
| 3518 | .set_height(1) |
| 3519 | .set_width(1) |
| 3520 | .set_layout(dnn::DataLayout::kBatchYXDepth); |
| 3521 | ScopedTensorDescriptor bias_descriptor{bias_dimensions, miopenFloat}; |
| 3522 | |
| 3523 | if (input_data.opaque() != output_data->opaque()) { |
| 3524 | stream->ThenMemcpy(output_data, input_data, |
| 3525 | dimensions.ElementCount() * sizeof(float)); |
| 3526 | if (!stream->ok()) { |
| 3527 | LOG(ERROR) |
| 3528 | << "stream " << stream |
| 3529 | << " could not enqueue a tensor copy as part of bias addition."; |
| 3530 | return false; |
| 3531 | } |
| 3532 | } |
| 3533 | |
| 3534 | auto miopen = miopen_->GetHandle(parent_, stream); |
| 3535 | |
| 3536 | const float alpha1 = 1.0f; |
| 3537 | const float alpha2 = 0.0f; |
| 3538 | const float beta = 1.0f; |
| 3539 | |
| 3540 | auto status = wrap::miopenOpTensor( |
| 3541 | miopen.handle(), miopenTensorOpAdd, &alpha1, bias_descriptor.handle(), |
| 3542 | biases.opaque(), &alpha2, bias_descriptor.handle(), biases.opaque(), |
| 3543 | &beta, input_descriptor.handle(), output_data->opaque()); |
| 3544 | |
| 3545 | if (status != miopenStatusSuccess) { |
| 3546 | LOG(ERROR) << "stream " << stream << " could not enqueue bias addition."; |
| 3547 | return false; |
| 3548 | } |
| 3549 | |
| 3550 | return true; |
| 3551 | } |
| 3552 | |
| 3553 | bool MIOpenSupport::DoActivate(Stream* stream, |
| 3554 | dnn::ActivationMode activation_mode, |
no test coverage detected