macro parameter won't take argument passed (nvcc)

Question

I'm just starting to code on CUDA and I'm trying to manage my codes into a bunch of different files, but one of my macros won't take the argument passed for some reason.

The error is:

addkernel.cu(19): error: identifier "err" is undefined

so my main code is in ../cbe4/addkernel.cu

#include <stdio.h>
#include <stdlib.h>

#include "cbe4.h"
#include "../mycommon/general.h"

#define N 100

int main( int argc, char ** argv ){

        float h_out[N], h_a[N], h_b[N]; 
        float *d_out, *d_a, *d_b; 

        for (int i=0; i<N; i++) {
                h_a[i] = i + 5;
                h_b[i] = i - 10;
        }

        // The error is on the next line
        CUDA_ERROR( cudaMalloc( (void **) &d_out, sizeof(float) * N ) );
        CUDA_ERROR( cudaMalloc( (void **) &d_a, sizeof(float) * N ) ); 
        CUDA_ERROR( cudaMalloc( (void **) &d_b, sizeof(float) * N ) );

        cudaFree(d_a);
        cudaFree(d_b);


        return EXIT_SUCCESS;
}

The macro is defined in ../mycommon/general.h:

#ifndef __GENERAL_H__
#define __GENERAL_H__

#include <stdio.h>

// error checking 
void CudaErrorCheck (cudaError_t err, const char *file, int line);

#define CUDA_ERROR ( err ) (CudaErrorCheck( err, __FILE__, __LINE__ )) 

#endif

and this is the source code for the function CudaErrorCheck in ../mycommon/general.cu:

#include <stdio.h>
#include <stdlib.h>

#include "general.h"

void CudaErrorCheck (cudaError_t err,
                        const char *file,
                        int line) {
        if ( err != cudaSuccess ) {
                printf( "%s in %s at line %d \n",
                        cudaGetErrorString( err ),
                        file, line );
                exit( EXIT_FAILURE );
        }
}

../cbe/cbe4.h is my header file and ../cbe/cbe4.cu the source file for kernel codes (in case this might help):

in cbe4.h:

__global__
void add( float *, float *, float * );

in cbe4.cu:

    #include "cbe4.h"

__global__ void add( float *d_out, float *d_a, float *d_b ) {
        int tid = (blockIdx.x * blockDim.x) + threadIdx.x;
        d_out[tid] = d_a[tid] + d_b[tid]; }

and here's my makefile (stored in ../cbe4):

NVCC = nvcc
SRCS = addkernel.cu cbe4.cu
HSCS = ../mycommon/general.cu

addkernel:  
        $(NVCC) $(SRCS) $(HSCS) -o $@

Also, I'm using the Cuda by Example book, by the way. One thing about the code in common/book.h, the function for HandleError ( I renamed it CudaErrorCheck and placed it in another source code here ) was defined in the header file (equivalently, at the CudaErrorCheck declaration in my general.h . Isn't this inadvisable? Or so I heard. )

Tangential: Note that you shouldn't create function or variable names that start with an underscore, in general. [C11 §7.1.3 Reserved identifiers](https://port70.net/~nsz/c/c11/n1570.html#7.1.3) says: — _All identifiers that begin with an underscore and either an uppercase letter or another underscore are always reserved for any use._ — _All identifiers that begin with an underscore are always reserved for use as identifiers with file scope in both the ordinary and tag name spaces._ See also [What does double underscore (`__const`) mean in C?](https://stackoverflow.com/a/1449301/15168) — Jonathan Leffler, Feb 07 '18 at 15:58
The error message is from compiling `addkernel.cu` — but that's one chunk of code that is missing from the question. We can't help you debug code we can't see. — Jonathan Leffler, Feb 07 '18 at 16:00
@JonathanLeffler what are you referring to specifically regarding the underscore? `__global__` is CUDA-specific syntax. It wasn't created by OP. — Robert Crovella, Feb 07 '18 at 16:01
@RobertCrovella: I'm referring primarily to `#ifndef __GENERAL_H__` in the user's own `general.h` header. If `__global__` is CUDA-specific syntax, it is exempt — it is defined by the implementation and the user isn't creating it, as you say. — Jonathan Leffler, Feb 07 '18 at 16:02
in that case I believe OP is following an example from a set of header files that are provided for code examples in the Cuda By Example book. [Here](https://github.com/jiekebo/CUDA-By-Example/blob/master/common/book.h#L17) is one such example. I acknowledge your stated concern, of course. However the blame probably equally rests with the book's authors, perhaps. — Robert Crovella, Feb 07 '18 at 16:13
@JonathanLeffler Robert is correct. I thought it was a convention. The main code is in addkernel.cu, I have mistakenly named it and I have corrected it. — Sunny Hill, Feb 07 '18 at 23:25
@Aram If I put all the codes in one file, there would be no problems compiling the code. I haven't been using any cuda headers so far. Coding this on a linux machine btw. — Sunny Hill, Feb 07 '18 at 23:26
The best I can do is point out that the C standard says that the advice in the book is dubious. You can ignore me and the standard and I won't be offended, and you'll most likely get away with, right up until you don't and then you'll have no recourse but to change your code because it violates the standard. This will usually happen when it is most inconvenient. — Jonathan Leffler, Feb 08 '18 at 01:45
@JonathanLeffler Actually I find it strange too. I will not adopt it from now on. — Sunny Hill, Feb 08 '18 at 02:06
Please don't add your solution into your question. Post it as an answer. — Robert Columbia, Feb 08 '18 at 02:07
FWIW (not much), I typically use `#ifndef GENERAL_H_INCLUDED` at the start of a header `general.h`. I've seen quite a variety of alternative notations (and a lot of notations that break the rules of the standard), including `GENERAL_DOT_H` or even `GENERAL_DOT_H_INCLUDED`. If you're really concerned about name collisions, consider using an MD5 hash of the header file contents. — Jonathan Leffler, Feb 08 '18 at 02:10
@JonathanLeffler I have have to ask. Any recommendations on great resources to improve my C programmming skills? I'm currently using Lu's Intermediate Programming in C book. — Sunny Hill, Feb 08 '18 at 02:17
There's a fairly good [tag:c] tag on Stack Overflow for general C question. There's the [Definitive C Book List and Guide](https://stackoverflow.com/questions/562303/the-definitive-c-book-guide-and-list) question specifically. There are numerous good Unix/POSIX programming books that might be relevant. Once you delve into Arduino and CUDA, you're moving outside my realm of expertise — and beware asking on SO about that as they're external resources and hence off-topic. — Jonathan Leffler, Feb 08 '18 at 17:11

Jonathan Leffler · Accepted Answer · 2018-02-08T02:46:50.210

Spacing matters in macro definitions. You have:

#define CUDA_ERROR ( err ) (CudaErrorCheck( err, __FILE__, __LINE__ ))

You need (minimal change — delete one space):

#define CUDA_ERROR( err ) (CudaErrorCheck( err, __FILE__, __LINE__ ))

With a function-like macro, there cannot be white space between the macro name and the open parenthesis of the argument list of the macro definition. When it comes to using the macro, white space is allowed between the macro name and the open parenthesis of the argument list.

I'd write:

#define CUDA_ERROR(err) CudaErrorCheck(err, __FILE__, __LINE__)

The extra parentheses around the whole expansion aren't really necessary, and I don't much like white space around parentheses. Different people have different views on this, so I'm stating my preference without in any sense demanding that you use it (but obviously suggesting that you consider it).

Because of the space, your code was expanding to look like:

( err ) (CudaErrorCheck( err, "addkernel.cu", 19 ))( cudaMalloc( (void **) &d_out, sizeof(float) * N ) );

and err was diagnosed as an undefined identifier, making the cast invalid.

Thanks so much! This is what I'm looking for. – Sunny Hill Feb 08 '18 at 02:02 — Sunny Hill, Feb 08 '18 at 02:02

macro parameter won't take argument passed (nvcc)

1 Answers1