Gemm kernel does not exist

[shidong-ai]

Hello, I was using MIOpen as an external library when implementing my own program. But below error occurs when I call the convolution forward API. I tried to link the miopengemm to the program, but it still doesn't work. Anyone got any clue? Thanks in advance!

ROCmSoftwarePlatform/MIOpen/src/gemm.cpp:329: looking for gemm kernel (does not exist): miopenConvolutionFwdAlgoGEMM, tC0_tA0_tB0_colMaj1_m65536_n32_k75_lda65536_ldb75_ldc65536_ws0_f32

[dagamayank]

@doody1986 Do you have miopengemm installed on your system, as noted in the Prerequisites section of the README?

If yes, can you set export MIOPEN_ENABLE_LOGGING=1 and provide the output.

[shidong-ai]

@dagamayank
I did install that, and the output is as below. Thanks a lot !
miopenStatus_t miopenCreateTensorDescriptor(miopenTensorDescriptor_t *){
tensorDesc = 0x6572695720646e61
}
miopenStatus_t miopenCreateTensorDescriptor(miopenTensorDescriptor_t *){
tensorDesc = 0x30363237
}
miopenStatus_t miopenCreateConvolutionDescriptor(miopenConvolutionDescriptor_t *){
convDesc = 0
}
miopenStatus_t miopenCreateTensorDescriptor(miopenTensorDescriptor_t *){
tensorDesc = 0x303632
}
miopenStatus_t miopenSet4dTensorDescriptor(miopenTensorDescriptor_t, miopenDataType_t, int, int, int, int){
tensorDesc =
dataType = 1
n = 100
c = 3
h = 256
w = 256
}
miopenStatus_t miopenInitConvolutionDescriptor(miopenConvolutionDescriptor_t, miopenConvolutionMode_t, int, int, int, int, int, int){
convDesc = 0, 0, 1, 1, 1, 1,
mode = 0
pad_h = 2
pad_w = 2
u = 1
v = 1
dilation_h = 1
dilation_w = 1
}
miopenStatus_t miopenSet4dTensorDescriptor(miopenTensorDescriptor_t, miopenDataType_t, int, int, int, int){
tensorDesc =
dataType = 1
n = 32
c = 3
h = 5
w = 5
}
miopenStatus_t miopenSet4dTensorDescriptor(miopenTensorDescriptor_t, miopenDataType_t, int, int, int, int){
tensorDesc =
dataType = 1
n = 100
c = 32
h = 256
w = 256
}
miopenStatus_t miopenConvolutionForwardGetWorkSpaceSize(miopenHandle_t, const miopenTensorDescriptor_t, const miopenTensorDescriptor_t, const miopenConvolutionDescriptor_t, const miopenTensorDescriptor_t, size_t *){
wDesc = 32, 3, 5, 5
yDesc = 100, 32, 256, 256
convDesc = 2, 2, 1, 1, 1, 1,
workSpaceSize = 3159602
}
miopenStatus_t miopenConvolutionBackwardWeightsGetWorkSpaceSize(miopenHandle_t, const miopenTensorDescriptor_t, const miopenTensorDescriptor_t, const miopenConvolutionDescriptor_t, const miopenTensorDescriptor_t, size_t *){
dyDesc = 100, 32, 256, 256
xDesc = 100, 3, 256, 256
convDesc = 2, 2, 1, 1, 1, 1,
dwDesc = 32, 3, 5, 5
workSpaceSize = 14416760
}
miopenStatus_t miopenConvolutionBackwardDataGetWorkSpaceSize(miopenHandle_t, const miopenTensorDescriptor_t, const miopenTensorDescriptor_t, const miopenConvolutionDescriptor_t, const miopenTensorDescriptor_t, size_t *){
dyDesc = 100, 32, 256, 256
wDesc = 32, 3, 5, 5
convDesc = 2, 2, 1, 1, 1, 1,
dxDesc = 100, 3, 256, 256
workSpaceSize = 0
}
miopenStatus_t miopenConvolutionForward(miopenHandle_t, const void *, const miopenTensorDescriptor_t, const void *, const miopenTensorDescriptor_t, const void *, const miopenConvolutionDescriptor_t, miopenConvFwdAlgorithm_t, const void *, const miopenTensorDescriptor_t, void *, void *, size_t){
alpha = 0x7f55c0a81138
xDesc = 100, 3, 256, 256
x = 0x22010a7000
wDesc = 32, 3, 5, 5
w = 0x226e6ab000
convDesc = 2, 2, 1, 1, 1, 1,
algo = 0
beta = 0x7f55c0a82a34
yDesc = 100, 32, 256, 256
y = 0x220a6a9000
workSpace = 0x226e6b3000
workSpaceSize = 19660800
}
MIOpen Error: /home/yifan/dev/src/github.com/ROCmSoftwarePlatform/MIOpen/src/gemm.cpp:329: looking for gemm kernel (does not exist): miopenConvolutionFwdAlgoGEMM, tC0_tA0_tB0_colMaj1_m65536_n32_k75_lda65536_ldb75_ldc65536_ws0_f32

[dagamayank]

How did you install miopengemm? The above convolution is working fine on my end.

[shidong-ai]

@dagamayank
I just used "sudo apt-get install miopengemm. Is that all right?

[shidong-ai]

@dagamayank
I tried to build an MIOpen library with miopengemm instead of using the one installed, but it still doesn't work.

[patflick]

@doody1986

You have to run miopenFindConvolutionForwardAlgorithm(...) prior to miopenConvolutionForward.

In your log above it looks like you're not doing this. The Find type functions will create and compile multiple kernels, and then run each one once to find the fastest one. The miopenConvolutionForward function then just tries to load the kernel, but fails because the kernel was not yet created/compiled.

[shidong-ai]

Thanks a lot! That explains a lot. I thought it would work when I hard-coded the algorithm in my code

[dagamayank]

@patflick Thanks for catching that. I also added a wiki page to clarify a bit about MIOpen Convolutions.