cancel
Showing results for 
Search instead for 
Did you mean: 

Archives Discussions

spectral
Adept II

Compilation error : "declared implicitly"

Hi,

In my OpenCL code I need to call a method that will do a specific processing on a "set of datas".

So I have create a method like this :

int Intersects_Primitive(float * datas, IntersectionResult * result)

{

...

}

 

But when I try to call it, I get a compilation error !! I'm sure that there is a reason, but which one ? and what solution to this problem ?

Intersects_Primitive(&datas, &result)

I got the following message "declared implicitly"... what does it mean ?

Thanks

0 Likes
3 Replies
genaganna
Journeyman III

could you please post kernel code to reproduce this error so that we can look for workaround?

0 Likes

Hi,

Here is my code :

 

typedef struct

{

/// <summary>

/// The bounding box's coordinates

/// </summary>

float MinX, MinY, MinZ, MaxX, MaxY, MaxZ;

} AABB;

typedef struct

{

float Maximum;

int PrimitiveId;

float U;

float V;

} IntersectionResult;

typedef struct

{

float x, y, z, w;

} Vector3;

typedef struct

{

/// <summary>

/// The bounding box's coordinates

/// </summary>

AABB BBox;

/// <summary>

/// The index of the sibbling node (next node at the same level).

/// </summary>

int SkipNodeIndex;

/// <summary>

/// The primitive Id.

/// </summary>

/// <remarks>If it is a set (Count > -1) it is the first index in the primitives' ids list (_primitivesIds)</remarks>

int PrimitiveId;

/// <summary>

/// The number of primitives, used when it is a set.

/// </summary>

ushort Count;

} BoundingVolume;

kernel void TreeTraversal(

global Vector3 * origin,

global Vector3 * direction,

float rayMaximum,

int isShadowRay, // 1 = true, 0 = false

global Vector3 * rayOffset,

global BoundingVolume * _nodes,

global int * primitivesIds,

global Vector3 * primitivesVertices,

global IntersectionResult * result)

{

int OCLIndex = get_global_id(0);

 

int bvNodeIndex = 0;

//MITData mitData = new MITData(ray);

// End of the tree

int stopNodeIndex = _nodes[bvNodeIndex].SkipNodeIndex;

int hasIntersection = 0;

 

int loopCount = 0;

while (bvNodeIndex < stopNodeIndex)

{

loopCount++;

// Test for the best primitives

//if (loopCount < 1 /*&& hitsCache != null*/)

//{

// int[] bestHits = hitsCache.Primitives;

// if (bestHits != null)

// for (int index = 0; index < bestHits.Length; index++)

// _primitiveList.IntersectPrimitive(ray, bestHits[index], istate);

//}

 

// Do a hit test with the bounding volume.

// We use the "closest intersection" to check against the BV intersection.

// If the 'closest intersection' < 'bv intersection' then there is no

// primitive intersection possible !

float minHit;

float maxHit;

int hasHit = Intersects_BoxRay(/*mitData,*/ origin, direction, _nodes[bvNodeIndex].BBox, &minHit, &maxHit);

if (hasHit &&

minHit <= rayMaximum)

{

// It is a leaf -> test the primitives

if (_nodes[bvNodeIndex].PrimitiveId > -1)

{

origin = origin + rayOffset;

// Contains a simple primitive

if (_nodes[bvNodeIndex].Count < 1)

{

if (Intersects_Primitive(primitivesVertices, origin, direction, _nodes[bvNodeIndex].PrimitiveId, result))

{

// If shadow ray

if (isShadowRay > 0)

return true;

//if (hitsCache != null)

// hitsCache.AddHit(_nodes[bvNodeIndex].PrimitiveId, ray.Maximum);

hasIntersection = 1;

}

}

// Contains a set of primitive

else if (_nodes[bvNodeIndex].Count > 0)

{

int startIndex = bv.PrimitiveId;

int endIndex = startIndex + bv.Count - 1;

for (int index = startIndex; index <= endIndex; index++)

if (Intersects_Primitive(primitivesVertices, origin, direction, _primitivesIds[index], result))

{

// If shadow ray

if (isShadowRay > 0)

return true;

//if (hitsCache != null)

// hitsCache.AddHit(_primitivesIds[index], ray.Maximum);

hasIntersection = 1;

}

}

origin -= rayOffset;

}

// Next node at the same level OR

// the next sibbling of the parent.

bvNodeIndex++;

if ((bvNodeIndex >= stopNodeIndex || bvNodeIndex == _nodes[bvNodeIndex].SkipNodeIndex) &&

hasIntersection)

return 1;

}

// Continue at the same level

else

bvNodeIndex = _nodes[bvNodeIndex].SkipNodeIndex;

}

return 0;

}

int Intersects_BoxRay(struct Vector3 * origin, struct Vector3 * direction, struct BoundingVolume * aabb, float * minHit, float * maxHit)

{

return 0;

}

int Intersects_Primitive(float * primitivesVertices, struct Vector3 * origin, struct Vector3 * direction, int * _primitivesIds, struct IntersectionResult * result)

{

return 0;

}



0 Likes

OpenCL is based on c99 with some extensions and restrictions.

 

1. functions should be defined before invoking

2. pointer casting has lot of restrictions. see 6.8 section in OpenCL spec

3. user is responsible for arithmatic for user defined variables.

4. There are restrictions on structure alignments. see release notes

 

I removed all errors from your code. see attached code.

  Note : Initialized dummy values for undefined variables.

typedef struct { /// <summary> /// The bounding box's coordinates /// </summary> float MinX, MinY, MinZ, MaxX, MaxY, MaxZ; } AABB; typedef struct { float Maximum; int PrimitiveId; float U; float V; }IntersectionResult; typedef struct { float x, y, z, w; } Vector3; typedef struct { /// <summary> /// The bounding box's coordinates /// </summary> AABB BBox; /// <summary> /// The index of the sibbling node (next node at the same level). /// </summary> int SkipNodeIndex; /// <summary> /// The primitive Id. /// </summary> /// <remarks>If it is a set (Count > -1) it is the first index in the primitives' ids list (_primitivesIds)</remarks> int PrimitiveId; /// <summary> /// The number of primitives, used when it is a set. /// </summary> ushort Count; }BoundingVolume; int Intersects_BoxRay(global Vector3 * origin, global Vector3 * direction, global AABB* aabb, float * minHit, float * maxHit) { return 0; } int Intersects_Primitive(global Vector3 * primitivesVertices, global Vector3 * origin, global Vector3 * direction, int _primitivesIds, global IntersectionResult * result) { return 0; } kernel void TreeTraversal( global Vector3 * origin, global Vector3 * direction, float rayMaximum, int isShadowRay, // 1 = true, 0 = false global Vector3 * rayOffset, global BoundingVolume * _nodes, global int * primitivesIds, global Vector3 * primitivesVertices, global IntersectionResult * result) { int OCLIndex = get_global_id(0); int bvNodeIndex = 0; //MITData mitData = new MITData(ray); // End of the tree int stopNodeIndex = _nodes[bvNodeIndex].SkipNodeIndex; int hasIntersection = 0; int loopCount = 0; while (bvNodeIndex < stopNodeIndex) { loopCount++; // Test for the best primitives //if (loopCount < 1 /*&& hitsCache != null*/) //{ // int[] bestHits = hitsCache.Primitives; // if (bestHits != null) // for (int index = 0; index < bestHits.Length; index++) // _primitiveList.IntersectPrimitive(ray, bestHits[index], istate); //} // Do a hit test with the bounding volume. // We use the "closest intersection" to check against the BV intersection. // If the 'closest intersection' < 'bv intersection' then there is no // primitive intersection possible ! float minHit; float maxHit; int hasHit = Intersects_BoxRay(/*mitData,*/ origin, direction, &_nodes[bvNodeIndex].BBox, &minHit, &maxHit); if (hasHit && minHit <= rayMaximum) { // It is a leaf -> test the primitives if (_nodes[bvNodeIndex].PrimitiveId > -1) { origin[0].x = origin[0].x + rayOffset[0].x; origin[0].y = origin[0].y + rayOffset[0].y; origin[0].z = origin[0].z + rayOffset[0].z; //origin = origin + rayOffset; // Contains a simple primitive if(_nodes[bvNodeIndex].Count < 1) { if(Intersects_Primitive(primitivesVertices, origin, direction, _nodes[bvNodeIndex].PrimitiveId, result)) { // If shadow ray if (isShadowRay > 0) return; //if (hitsCache != null) // hitsCache.AddHit(_nodes[bvNodeIndex].PrimitiveId, ray.Maximum); hasIntersection = 1; } } // Contains a set of primitive else if (_nodes[bvNodeIndex].Count > 0) { //Note : bv variable not defined int startIndex = 0;//bv.PrimitiveId; int endIndex = startIndex + 0;//bv.Count - 1; for (int index = startIndex; index <= endIndex; index++) if(Intersects_Primitive(primitivesVertices, origin, direction, primitivesIds[index], result)) { // If shadow ray if (isShadowRay > 0) return; //if (hitsCache != null) // hitsCache.AddHit(_primitivesIds[index], ray.Maximum); hasIntersection = 1; } } //Note : Arithmetic on user defined variables //origin[0] -= rayOffset origin[0].x -= rayOffset[0].x; origin[0].y -= rayOffset[0].y; origin[0].z -= rayOffset[0].z; } // Next node at the same level OR // the next sibbling of the parent. bvNodeIndex++; if ((bvNodeIndex >= stopNodeIndex || bvNodeIndex == _nodes[bvNodeIndex].SkipNodeIndex) && hasIntersection) return; } // Continue at the same level else bvNodeIndex = _nodes[bvNodeIndex].SkipNodeIndex; } return; }

0 Likes