Skip to content
GitLab
Menu
Projects
Groups
Snippets
Loading...
Help
Help
Support
Community forum
Keyboard shortcuts
?
Submit feedback
Contribute to GitLab
Sign in / Register
Toggle navigation
Menu
Open sidebar
gaoqiong
composable_kernel
Commits
91669fc0
"examples/vscode:/vscode.git/clone" did not exist on "b98b314b7aa1b95829b316fb58aa9cabbb6fd2a6"
Commit
91669fc0
authored
May 17, 2023
by
rocking
Browse files
Fix typo
parent
6f2ef102
Changes
4
Hide whitespace changes
Inline
Side-by-side
Showing
4 changed files
with
17 additions
and
17 deletions
+17
-17
include/ck/tensor_operation/gpu/device/device_pool_fwd.hpp
include/ck/tensor_operation/gpu/device/device_pool_fwd.hpp
+1
-1
include/ck/tensor_operation/gpu/device/impl/device_pool2d_fwd_nhwc_nhwc.hpp
...operation/gpu/device/impl/device_pool2d_fwd_nhwc_nhwc.hpp
+4
-4
include/ck/tensor_operation/gpu/device/impl/device_pool3d_fwd_ndhwc_ndhwc.hpp
...eration/gpu/device/impl/device_pool3d_fwd_ndhwc_ndhwc.hpp
+4
-4
library/src/tensor_operation_instance/gpu/pool_fwd/pool_fwd_instance_common.hpp
...ration_instance/gpu/pool_fwd/pool_fwd_instance_common.hpp
+8
-8
No files found.
include/ck/tensor_operation/gpu/device/device_pool_fwd.hpp
View file @
91669fc0
...
@@ -19,7 +19,7 @@ template <index_t InOutRank,
...
@@ -19,7 +19,7 @@ template <index_t InOutRank,
typename
OutDataType
,
typename
OutDataType
,
typename
IndexDataType
,
typename
IndexDataType
,
ReduceTensorOp
ReduceOpId
,
ReduceTensorOp
ReduceOpId
,
bool
OuputIndex
>
bool
Ou
t
putIndex
>
struct
DevicePoolFwd
:
public
BaseOperator
struct
DevicePoolFwd
:
public
BaseOperator
{
{
virtual
std
::
unique_ptr
<
BaseArgument
>
virtual
std
::
unique_ptr
<
BaseArgument
>
...
...
include/ck/tensor_operation/gpu/device/impl/device_pool2d_fwd_nhwc_nhwc.hpp
View file @
91669fc0
...
@@ -20,10 +20,10 @@ namespace device {
...
@@ -20,10 +20,10 @@ namespace device {
template
<
typename
InDataType
,
template
<
typename
InDataType
,
typename
OutDataType
,
typename
OutDataType
,
typename
IndexDataType
,
// enable if OuputIndex == true
typename
IndexDataType
,
// enable if Ou
t
putIndex == true
typename
AccDataType
,
typename
AccDataType
,
ck
::
ReduceTensorOp
ReduceOpId
,
ck
::
ReduceTensorOp
ReduceOpId
,
bool
OuputIndex
,
bool
Ou
t
putIndex
,
ck
::
index_t
BlockSize
,
ck
::
index_t
BlockSize
,
ck
::
index_t
ReduceMThreadClusterSize
,
ck
::
index_t
ReduceMThreadClusterSize
,
ck
::
index_t
ReduceKThreadClusterSize
,
ck
::
index_t
ReduceKThreadClusterSize
,
...
@@ -31,7 +31,7 @@ template <typename InDataType,
...
@@ -31,7 +31,7 @@ template <typename InDataType,
ck
::
index_t
ReduceKThreadSliceSize
,
ck
::
index_t
ReduceKThreadSliceSize
,
ck
::
index_t
InSrcOutDstVectorSize
>
ck
::
index_t
InSrcOutDstVectorSize
>
struct
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
struct
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
:
public
DevicePoolFwd
<
4
,
2
,
InDataType
,
OutDataType
,
IndexDataType
,
ReduceOpId
,
OuputIndex
>
:
public
DevicePoolFwd
<
4
,
2
,
InDataType
,
OutDataType
,
IndexDataType
,
ReduceOpId
,
Ou
t
putIndex
>
{
{
static
constexpr
auto
I0
=
Number
<
0
>
{};
static
constexpr
auto
I0
=
Number
<
0
>
{};
static
constexpr
auto
I1
=
Number
<
1
>
{};
static
constexpr
auto
I1
=
Number
<
1
>
{};
...
@@ -230,7 +230,7 @@ struct DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
...
@@ -230,7 +230,7 @@ struct DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
const
auto
kernel
=
const
auto
kernel
=
kernel_reduce_threadwise
<
gridwise_reduce
,
kernel_reduce_threadwise
<
gridwise_reduce
,
OuputIndex
,
Ou
t
putIndex
,
true
,
// pooling need to return global index
true
,
// pooling need to return global index
false
,
// don't have index input
false
,
// don't have index input
InDataType
,
InDataType
,
...
...
include/ck/tensor_operation/gpu/device/impl/device_pool3d_fwd_ndhwc_ndhwc.hpp
View file @
91669fc0
...
@@ -20,10 +20,10 @@ namespace device {
...
@@ -20,10 +20,10 @@ namespace device {
template
<
typename
InDataType
,
template
<
typename
InDataType
,
typename
OutDataType
,
typename
OutDataType
,
typename
IndexDataType
,
// enable if OuputIndex == true
typename
IndexDataType
,
// enable if Ou
t
putIndex == true
typename
AccDataType
,
typename
AccDataType
,
ck
::
ReduceTensorOp
ReduceOpId
,
ck
::
ReduceTensorOp
ReduceOpId
,
bool
OuputIndex
,
bool
Ou
t
putIndex
,
ck
::
index_t
BlockSize
,
ck
::
index_t
BlockSize
,
ck
::
index_t
MThreadClusterSize
,
ck
::
index_t
MThreadClusterSize
,
ck
::
index_t
KThreadClusterSize
,
ck
::
index_t
KThreadClusterSize
,
...
@@ -31,7 +31,7 @@ template <typename InDataType,
...
@@ -31,7 +31,7 @@ template <typename InDataType,
ck
::
index_t
KThreadSliceSize
,
ck
::
index_t
KThreadSliceSize
,
ck
::
index_t
InSrcOutDstVectorSize
>
ck
::
index_t
InSrcOutDstVectorSize
>
struct
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
struct
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
:
public
DevicePoolFwd
<
5
,
3
,
InDataType
,
OutDataType
,
IndexDataType
,
ReduceOpId
,
OuputIndex
>
:
public
DevicePoolFwd
<
5
,
3
,
InDataType
,
OutDataType
,
IndexDataType
,
ReduceOpId
,
Ou
t
putIndex
>
{
{
static
constexpr
auto
I0
=
Number
<
0
>
{};
static
constexpr
auto
I0
=
Number
<
0
>
{};
static
constexpr
auto
I1
=
Number
<
1
>
{};
static
constexpr
auto
I1
=
Number
<
1
>
{};
...
@@ -235,7 +235,7 @@ struct DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
...
@@ -235,7 +235,7 @@ struct DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
const
auto
kernel
=
const
auto
kernel
=
kernel_reduce_threadwise
<
gridwise_reduce
,
kernel_reduce_threadwise
<
gridwise_reduce
,
OuputIndex
,
Ou
t
putIndex
,
true
,
// pooling need to return global index
true
,
// pooling need to return global index
false
,
// don't have index input
false
,
// don't have index input
InDataType
,
InDataType
,
...
...
library/src/tensor_operation_instance/gpu/pool_fwd/pool_fwd_instance_common.hpp
View file @
91669fc0
...
@@ -24,13 +24,13 @@ template <typename InDataType,
...
@@ -24,13 +24,13 @@ template <typename InDataType,
typename
IndexDataType
,
typename
IndexDataType
,
typename
AccDataType
,
typename
AccDataType
,
ReduceTensorOp
ReduceOpId
,
ReduceTensorOp
ReduceOpId
,
bool
OuputIndex
>
bool
Ou
t
putIndex
>
using
device_pooling2d_fwd_nhwc_instances
=
using
device_pooling2d_fwd_nhwc_instances
=
// clang-format off
// clang-format off
std
::
tuple
<
std
::
tuple
<
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
OuputIndex
,
256
,
256
,
1
,
1
,
1
,
1
>
,
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
Ou
t
putIndex
,
256
,
256
,
1
,
1
,
1
,
1
>
,
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
OuputIndex
,
256
,
256
,
1
,
2
,
1
,
2
>
,
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
Ou
t
putIndex
,
256
,
256
,
1
,
2
,
1
,
2
>
,
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
OuputIndex
,
256
,
256
,
1
,
4
,
1
,
4
>
DevicePool2dFwd_Input_N_Hi_Wi_C_Output_N_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
Ou
t
putIndex
,
256
,
256
,
1
,
4
,
1
,
4
>
// clang-format on
// clang-format on
>
;
>
;
...
@@ -39,13 +39,13 @@ template <typename InDataType,
...
@@ -39,13 +39,13 @@ template <typename InDataType,
typename
IndexDataType
,
typename
IndexDataType
,
typename
AccDataType
,
typename
AccDataType
,
ReduceTensorOp
ReduceOpId
,
ReduceTensorOp
ReduceOpId
,
bool
OuputIndex
>
bool
Ou
t
putIndex
>
using
device_pooling3d_fwd_ndhwc_instances
=
using
device_pooling3d_fwd_ndhwc_instances
=
// clang-format off
// clang-format off
std
::
tuple
<
std
::
tuple
<
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
OuputIndex
,
256
,
256
,
1
,
1
,
1
,
1
>
,
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
Ou
t
putIndex
,
256
,
256
,
1
,
1
,
1
,
1
>
,
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
OuputIndex
,
256
,
256
,
1
,
2
,
1
,
2
>
,
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
Ou
t
putIndex
,
256
,
256
,
1
,
2
,
1
,
2
>
,
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
OuputIndex
,
256
,
256
,
1
,
4
,
1
,
4
>
DevicePool3dFwd_Input_N_Di_Hi_Wi_C_Output_N_Do_Ho_Wo_C
<
InDataType
,
OutDataType
,
IndexDataType
,
AccDataType
,
ReduceOpId
,
Ou
t
putIndex
,
256
,
256
,
1
,
4
,
1
,
4
>
// clang-format on
// clang-format on
>
;
>
;
...
...
Write
Preview
Markdown
is supported
0%
Try again
or
attach a new file
.
Attach a file
Cancel
You are about to add
0
people
to the discussion. Proceed with caution.
Finish editing this message first!
Cancel
Please
register
or
sign in
to comment