cudagpu-constant-memory

constant memory


I write code for copying an integer to constant memory and use it in a global function,but it doesn't work correctly.I mean that no "cuPrintf" in global function works and nothing printed!

I think it is because of "if(*num == 5)",since I remove it, all "cuPrintf" print what I want!

I also try "if(num == 5)" but my visual studio doesn't compile it and shows error.

why "num" is correct in "cuPrintf" but it is not correct in "if" statement?

how should I use "num" in an "if" statement?

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include "stdio.h"
#include "stdlib.h"
#include "cuPrintf.cu"

__constant__ int* num;

__global__ void kernel(){
cuPrintf("\n num=%d\n",num);
if(*num == 5)
    cuPrintf("\n num is equal 5");
else
    cuPrintf("\n num is not equal 5");
}

void main(){    
int x;
printf("\n\nPlease enter x:");
scanf("%d",&x);
cudaMemcpyToSymbol( num, &x,sizeof(int)*1);

cudaPrintfInit();
kernel<<<1,1>>>();
cudaPrintfDisplay(stdout, true);
cudaPrintfEnd();

int wait;
scanf("%d",&wait);
}

if I change:

__constant__ int* num;

to

__constant__ int num;

and also change:

cudaMemcpyToSymbol( num, &x,sizeof(int)*1);

to

cudaMemcpyToSymbol( &num, &x,sizeof(int)*1);

then

cuPrintf("\n num=%d\n",num);

will show "num=0" with any input!


Solution

  • "num" should not be a pointer. I changed your code to the one below, works for me (note it requires SM 2.0 or newer for printf):

    #include "cuda_runtime.h"
    #include "device_launch_parameters.h"
    
    #include "stdio.h"
    #include "stdlib.h"
    
    __constant__ int num;
    
    __global__ void kernel() {
        printf("\n num=%d\n", num);
        if (num == 5)
            printf("\n num is equal 5");
        else
            printf("\n num is not equal 5");
    }
    
    int main() {
        cudaError_t err;
        int x;
        printf("\n\nPlease enter x:");
        scanf("%d", &x);
        err = cudaMemcpyToSymbol(num, &x, sizeof(int) * 1);
        if (err != cudaSuccess) {
            printf("Error: %d\n", err);
            return 1;
        }
    
        kernel<<<1, 1>>>();
        err = cudaDeviceSynchronize();
        if (err != cudaSuccess) {
            printf("Error: %d\n", err);
            return 1;
        }
        return 0;
    }