can you use `restrict` on data members with cudaMalloc?

Accelerated Computing CUDA CUDA Programming and Performance
1

I am redoing some code to make sure I am taking advantage of restrict where applicable. For functions, it’s golden. But I’d like to also apply it to struct members, which is supposed to be valid: http://en.cppreference.com/w/c/language/restrict#Struct_members

However, I can’t figure out how to do this (or if it is even possible) because cudaMalloc needs a void **. Example

struct MyData {
    float4 * __restrict__ a = nullptr;
    float4 * __restrict__ b = nullptr;

    static const int N = 20;

    void allocate() {
        cudaMalloc(
            // reinterpret_cast<void **>(&a), // cannot cast away 'const' or other qualifiers
            &a,
            N * sizeof(float4)
        );
        cudaMalloc(
            &b,
            N * sizeof(float4)
        );            
    }

    void free() {
        cudaFree(a);
        cudaFree(b);
    }
};

int main(void) {
    MyData data;
    data.allocate();
    data.free();
}

Errors with

error: no instance of overloaded function "cudaMalloc" matches the argument list
            argument types are: (float4 *__restrict__ *, unsigned long)

In reality I think I have all of my bases covered, the methods where a and b are actually used in functions are marked as restrict. I was just wondering if the above is supported, planned to be, not possible to support, etc. Thanks for any thoughts!

2

this variant builds for me with nvcc -std=c++11 -o test test.cu from the either the CUDA 7.5, 8.0 or the 9.1 toolkit.

sometimes the simplest typecasts are the best ;)

void allocate() {
            cudaMalloc(
                (void**)&a,
                N * sizeof(float4)
            );
            cudaMalloc(
                (void**)&b,
                N * sizeof(float4)
            );            
        }
3

Indeed! I can’t believe I didn’t even try that faceslap. Thanks :)