ETRI IVCL 1.0.0
Acceleration SW Platform for Ondevice
All Classes Namespaces Files Functions Variables Typedefs Enumerations Enumerator Friends Macros Pages
Functions
bp.cl File Reference

Functions

__kernel void initializeHierCost (__global float *datacost, __global float *hierCost, int width, int height, int dispRange)
 
__kernel void initializeHierCost_dispSplit (__global float *datacost, __global float *hierCost, int width, int height, int dispRange)
 
void gpu_comp_msg (__global float *m1, __global float *m2, __global float *m3, __global float *data, __global float *dest, int MAX_DISP, float DISCONT_COST, float CONT_COST)
 
__kernel void hierBP (__global float *temp_m_u, __global float *temp_m_d, __global float *temp_m_l, __global float *temp_m_r, __global float *datacost, __global float *m_u, __global float *m_d, __global float *m_l, __global float *m_r, __global float *gpu_zero, int width, int height, int dispRange, int dist, float DISCONT_COST, float CONT_COST)
 
__kernel void updateCostLayer (__global float *m_u, __global float *m_d, __global float *m_l, __global float *m_r, int width, int height, int dispRange, int dist)
 
__kernel void finalBP (__global float *m_u, __global float *m_d, __global float *m_l, __global float *m_r, __global float *datacost, __global float *gpu_zero, int width, int height, int dispRange)
 
__kernel void updateCostLayer_dispSplit (__global float *m_u, __global float *m_d, __global float *m_l, __global float *m_r, int width, int height, int dispRange, int dist)
 
__kernel void finalBP_dispSplit (__global float *m_u, __global float *m_d, __global float *m_l, __global float *m_r, __global float *datacost, __global float *gpu_zero, int width, int height, int dispRange)
 

Function Documentation

◆ finalBP()

__kernel void finalBP ( __global float *  m_u,
__global float *  m_d,
__global float *  m_l,
__global float *  m_r,
__global float *  datacost,
__global float *  gpu_zero,
int  width,
int  height,
int  dispRange 
)

The final belief propagation in the base layer

Parameters
m_uThe array storage for node in up direction
m_dThe array storage for node in down direction
m_lThe array storage for node in left direction
m_rThe array storage for node in right direction
datacostThe 3D datacost
gpu_zeroArray containing zero to be used in GPU
widthThe image width
heightThe image height
dispRangeThe disparity search range

◆ finalBP_dispSplit()

__kernel void finalBP_dispSplit ( __global float *  m_u,
__global float *  m_d,
__global float *  m_l,
__global float *  m_r,
__global float *  datacost,
__global float *  gpu_zero,
int  width,
int  height,
int  dispRange 
)

The final belief propagation in the base layer
using dispSplit

Parameters
m_uThe array storage for node in up direction
m_dThe array storage for node in down direction
m_lThe array storage for node in left direction
m_rThe array storage for node in right direction
datacostThe 3D datacost
gpu_zeroArray containing zero to be used in GPU
widthThe image width
heightThe image height
dispRangeThe disparity search range

◆ gpu_comp_msg()

void gpu_comp_msg ( __global float *  m1,
__global float *  m2,
__global float *  m3,
__global float *  data,
__global float *  dest,
int  MAX_DISP,
float  DISCONT_COST,
float  CONT_COST 
)

Belief message propagation computation. This uses local memory (faster) if the GPU supports

Parameters
m1The 1st neighbor node
m2The 2nd neighbor node
m3The 3rd neighbor node
dataThe current node value
destThe node to store the compute belief message
MAX_DISPThe disparity search range
DISCONT_COSTThe edge hyperparameter
CONT_COSTThe smoothing hyperparameter

◆ hierBP()

__kernel void hierBP ( __global float *  temp_m_u,
__global float *  temp_m_d,
__global float *  temp_m_l,
__global float *  temp_m_r,
__global float *  datacost,
__global float *  m_u,
__global float *  m_d,
__global float *  m_l,
__global float *  m_r,
__global float *  gpu_zero,
int  width,
int  height,
int  dispRange,
int  dist,
float  DISCONT_COST,
float  CONT_COST 
)

The interface for hierarchal belief propagation using only global memory

Parameters
temp_m_uThe temporary array storage for node in up direction
temp_m_dThe temporary array storage for node in down direction
temp_m_lThe temporary array storage for node in left direction
temp_m_rThe temporary array storage for node in right direction
datacostThe 3D datacost
m_uThe array storage for node in up direction
m_dThe array storage for node in down direction
m_lThe array storage for node in left direction
m_rThe array storage for node in right direction
gpu_zeroArray containing zero to be used in GPU
widthThe image width
heightThe image height
dispRangeThe disparity search range
distThe distance value used for indexing
DISCONT_COSTThe edge hyperparameter
CONT_COSTThe smoothing hyperparameter

◆ initializeHierCost()

__kernel void initializeHierCost ( __global float *  datacost,
__global float *  hierCost,
int  width,
int  height,
int  dispRange 
)

Interpolate the initial cost (base) into a hierachal structure datacost

Parameters
datacostThe initial 3D cost
hierCostHierarchal 3D cost
widthThe image width
heightThe image height
dispRangeThe stereo disparity search range

◆ initializeHierCost_dispSplit()

__kernel void initializeHierCost_dispSplit ( __global float *  datacost,
__global float *  hierCost,
int  width,
int  height,
int  dispRange 
)

Interpolate the initial cost (base) into a hierachal structure datacost
when dispSplit

Parameters
datacostThe initial 3D cost
hierCostHierarchal 3D cost
widthThe image width
heightThe image height
dispRangeThe stereo disparity search range

◆ updateCostLayer()

__kernel void updateCostLayer ( __global float *  m_u,
__global float *  m_d,
__global float *  m_l,
__global float *  m_r,
int  width,
int  height,
int  dispRange,
int  dist 
)

Update the computed cost into the next hierarchal layer

Parameters
m_uThe array storage for node in up direction
m_dThe array storage for node in down direction
m_lThe array storage for node in left direction
m_rThe array storage for node in right direction
widthThe image width
heightThe image height
dispRangeThe disparity search range
distThe distance value used for indexing

◆ updateCostLayer_dispSplit()

__kernel void updateCostLayer_dispSplit ( __global float *  m_u,
__global float *  m_d,
__global float *  m_l,
__global float *  m_r,
int  width,
int  height,
int  dispRange,
int  dist 
)

Update the computed cost into the next hierarchal layer
using dispSplit

Parameters
m_uThe array storage for node in up direction
m_dThe array storage for node in down direction
m_lThe array storage for node in left direction
m_rThe array storage for node in right direction
widthThe image width
heightThe image height
dispRangeThe disparity search range
distThe distance value used for indexing