char \0 escape sequence - what does it mean?

Question

I'm reading this example

The real "Hello World!" for CUDA!
What does the \0 in

char str[16] = "Hello \0\0\0\0\0\0";

stand for?

I'm not sure why the 16 char str has "Hello " inside it and then all zeroes then (this isn't a global variable). How can I be sure that it just contains zeroes?"

Answer 1

'\0' is the ASCII NUL null character (ASCII code zero).

There is no point in specifying all of the zero bytes in the array. The following are equivalent:

char str[16] = "Hello \0\0\0\0\0\0";
char str[16] = "Hello ";

If an array is partially initialized, elements that are not initialized receive the value 0 of the appropriate type. ^[IBM]

Since the length was given to be 16, the compiler will automatically ensure the rest of the array (after "Hello " is zeroed.) The author did this either to "be sure" the rest of the array was zero, or to serve as documentation for the reader.

Contrary to our initial analyses, the CUDA kernel is not "poking in bytes" to append World! to the existing string. Hello is printed first. Then the kernel modifies the string to be World!, which is finally printed.

The only reason the string is specified to be 16 bytes, is because that is the block size the kernel is designed to work with, and they have to make sure that the kernel is not messing with memory it shouldn't be.

I've added some comments to the (partial) original code to make this all more clear:

__global__                              // The kernel which is run in parallel
void hello(char *a, int *b) 
{
    a[threadIdx.x] += b[threadIdx.x];
}

int main()
{
    // The line in question. There's really no point in padding it with zeros.
    // The zeros are *not* replaced, and only 12+1 bytes are being specified.
    char a[N] = "Hello \0\0\0\0\0\0";

    // These values are added (by the CUDA kernel) to the array above. Again,
    // since partial arrays are zero-filled, there's no point in filling this in.
    int b[N] = {15, 10, 6,  0, -11,  1, 0, 0, 0, 0, 0, 0, 0, 0, 0, 0};

    // 'H' + 15 = 'W'
    // 'e' + 10 = 'o'
    // 'l' + 6  = 'r'
    // 'l' + 0  = 'l'
    // 'o' - 11 = 'd'
    // ' ' + 1  = '!'

    char *ad;
    int *bd;
    const int csize = N*sizeof(char);
    const int isize = N*sizeof(int);

    printf("%s", a);                  // Print "Hello "

    cudaMalloc( (void**)&ad, csize ); 
    cudaMalloc( (void**)&bd, isize ); 
    cudaMemcpy( ad, a, csize, cudaMemcpyHostToDevice ); 
    cudaMemcpy( bd, b, isize, cudaMemcpyHostToDevice ); 

    dim3 dimBlock( blocksize, 1 );
    dim3 dimGrid( 1, 1 );
    hello<<<dimGrid, dimBlock>>>(ad, bd);   // Add the values in b to a


    cudaMemcpy( a, ad, csize, cudaMemcpyDeviceToHost ); 
    cudaFree( ad );
    cudaFree( bd );

    printf("%s\n", a);               // print "World!"
    return EXIT_SUCCESS;
}

Answer 2

\0 stands for NUL, Nul is used as a string termination character. Meaning it states the end of a string. The value of a NUL byte is 0x00

Answer 3

As stated it has little sense. \0 just places the character with code 0 there, but it happens anyway. The good use of this is where array bounds are not given, as the extra 0 will count in, or where you have more characters after the 0.

char foo_and_bar[] = "Foo\0Bar";

will separate the parts with 0.