Failure on assert

This is my sample code -

#include <cuda_runtime.h>
#include "helper_cuda.h"
#include "device_launch_parameters.h"
#include <iostream>

template <typename C, typename T>
__global__
void awkward_listarray_compact_offsets(T* tooffsets, const C* fromstarts, const C* fromstops, int64_t startsoffset, int64_t stopsoffset, int64_t length) {
	int idx = threadIdx.x + (blockIdx.x * blockDim.x);
  if (idx == 0) tooffsets[0] = 0;
  if (idx < length) {
    if (idx == 0) {
      for (int i = 0; i < length; i++) {
        C start = fromstarts[startsoffset + i];
        C stop = fromstops[stopsoffset + i];
        if (start < stop) assert(0);
        tooffsets[i + 1] = tooffsets[i] + (stop - start);
      }
    }
  }
}

template <typename T, typename C>
void offload(T* tooffsets, const C* fromstarts, const C* fromstops, int64_t startsoffset, int64_t stopsoffset, int64_t length) {
  int* d_tooffsets, * d_fromstarts, * d_fromstops;
  checkCudaErrors(cudaMalloc((void**)&d_tooffsets, (length + 1) * sizeof(int)));
  checkCudaErrors(cudaMalloc((void**)&d_fromstarts, length * sizeof(int)));
  checkCudaErrors(cudaMemcpy(d_fromstarts, fromstarts, length * sizeof(int), cudaMemcpyHostToDevice));
  checkCudaErrors(cudaMalloc((void**)&d_fromstops, length * sizeof(int)));
  checkCudaErrors(cudaMemcpy(d_fromstops, fromstops, length * sizeof(int), cudaMemcpyHostToDevice));
  int block, thread;
  if (length > 1024) {
    block = (length / 1024) + 1;
    thread = 1024;
  }
  else {
    thread = length;
    block = 1;
  }
  awkward_listarray_compact_offsets<int, int><<<block, thread>>>(d_tooffsets, d_fromstarts, d_fromstops, startsoffset, stopsoffset, length);
  checkCudaErrors(cudaDeviceSynchronize());
  checkCudaErrors(cudaMemcpy(tooffsets, d_tooffsets, (length + 1) * sizeof(int), cudaMemcpyDeviceToHost));
  checkCudaErrors(cudaFree(d_tooffsets));
  checkCudaErrors(cudaFree(d_fromstarts));
  checkCudaErrors(cudaFree(d_fromstops));
}

int main() {
  const int size = 100000;
  int tooffsets, fromstarts, fromstops;
  for (int i = 0; i < size; i++) {
    fromstarts[i] = i + 100;
    fromstops[i] = i + 10;
  }
  offload<int, int>(tooffsets, fromstarts, fromstops, 0, 0, size);
  for (int i = 0; i < size + 1; i++) {
    std::cout << tooffsets[i] << "\n";
  }
  return 0;
}

I am using checkCudaErrors() from the “helper_cuda.h” header file given in the NVIDIA CUDA SDK.

I am using the if (start<stop) block before asserting because my end goal is to call a function and then call assert to stop kernel execution and this is just me testing what works and what breaks and why.

If I compile the current code (which has assert(0) ), then I get the error -

Including cassert.h does not fix the issue.

But, when I compile the code with assert (start < stop), the code compiles and executes. But when start < stop condition is met, the assert function does not halt kernel execution and the code keeps executing.

What am I doing wrong?

There is no cassert.h

There is:

#include <assert.h>

or

#include

Either one of those should fix the compile error. Not including either one of those is definitely a coding error if you intend to use assert().

Thanks!

I still hit this issue. Could you please help me? :(
But, if I just use

assert(start < stop);

without the if conditional, then the assertion fails correctly.

Are you aware that

assert(start < stop);

and

if (start < stop) assert(0);

are logically opposite behaviors?

Yikes. Thanks.