Implementing Argmax and I

[Tanvi141]

Brief description of what is fixed or changed

Wrote code,tests,instantiation for Argmax. Code does not currently compile as I am working on figuring out passing function pointer to the kernel.

This is WIP at the time of making the pull request

[czgdp1807]

Please fix the following,

[==========] Running 1 test from 1 test suite.
[----------] Global test environment set-up.
[----------] 1 test from Cuda
[ RUN      ] Cuda.Argmax
/home/czgdp1807ssd/codezonediitj/adaboost/./adaboost/tests/test_cuda_operations.cu:37: Failure
Expected equality of these values:
  4
  result_gpu
    Which is: 0
The arg max value is at 4.
[  FAILED  ] Cuda.Argmax (209 ms)
[----------] 1 test from Cuda (209 ms total)

[----------] Global test environment tear-down
[==========] 1 test from 1 test suite ran. (209 ms total)
[  PASSED  ] 0 tests.
[  FAILED  ] 1 test, listed below:
[  FAILED  ] Cuda.Argmax

[Tanvi141]

Hi @czgdp1807, been trying to fix that error for a while now. The error is coming since func_ptr is not able to get correctly passed to the kernel. I followed this link: https://leimao.github.io/blog/Pass-Function-Pointers-to-Kernels-CUDA/#:~:text=The%20key%20to%20passing%20function,pointers%20to%20the%20host%20side

1. I think there is some problem coming with the passing of the __device__ pointer to Argmax. I tried to call CudaMemcpyFromSymbol and pass p_func to it. However I get error as invalid device symbol when I pass p_func.

2. I tried to implement the link as is, without passing pointer into the Argmax. I directly tried defining some random function and making a static pointer all within the same file. Here it worked perfectly.

3. So I thought maybe we can call cudaMemcpyFromSymbol in the test file to generate h_func variable, and then try passing it as an argument to Argmax. But for some reason it is not working here! The error I am getting is

   error: invalid conversion from ‘adaboost::cuda::core::func_t<float, float> {aka float (*)(float)}’ to ‘const void*’ 

   This error comes when I try to call cudaMemcpyFromSymbol in the test file. But the arguments I am passing are exactly similar to how they were passed in point 2.

I do not know why I am getting this error here. If you have any solutions/insights please do share.

[Tanvi141]

Hi @czgdp1807, I think I figured out a workaround to the problem. What we can do is create a separate file where the required functions and their pointers can be declared. Here I have done it in the file function_params.cu. Please see the latest commit. So the user will need to make changes in this file for using function parameters. Will this be okay?

Note: Currently the first argument in the Argmax code is redundant, if the above implementation is okay, then I will make the change.

[czgdp1807]

I will test the approach locally, will push some changes to your branch. Please enable, Allow edits from maintainers on all your PRs.

[Tanvi141]

It is done for both

[czgdp1807]

It is ready to go. Just needs some minor changes in other modules.

[czgdp1807]

Now, I think, Sum needs to be added along similar lines?