CIVL: Formal Verification of Parallel Programs

Zheng, Ming; Rogers, Michael S.; Luo, Ziqing; Dwyer, Matthew B.; Siegel, Stephen F.

doi:10.1109/ase.2015.99

Cited by 18 publications

(25 citation statements)

References 15 publications

Supporting

Mentioning

Contrasting

Order By: Relevance

“…As GPUVerify, PUG does not verify main functions and this explains most unsupported cases (31), while others are explained by the lack of support to __syncthreads function (12), function pointers (9), and the curand.h library (7); additionally, PUG does not support the use of unsigned type modifier as argument to the function atomi-cAdd (6); changes in variables stored in constant memory (3), and inability to handle structs (2), variables with __device__ qualifier (2), and size_t type (1), in addition to other cases that PUG aborted by returning a false null pointer access (7) or because it did not recognize the NULL identifier (2). As GPUVerify, PUG does not verify main functions and this explains most unsupported cases (31), while others are explained by the lack of support to __syncthreads function (12), function pointers (9), and the curand.h library (7); additionally, PUG does not support the use of unsigned type modifier as argument to the function atomi-cAdd (6); changes in variables stored in constant memory (3), and inability to handle structs (2), variables with __device__ qualifier (2), and size_t type (1), in addition to other cases that PUG aborted by returning a false null pointer access (7) or because it did not recognize the NULL identifier (2).…”

Section: Resultsmentioning

confidence: 99%

“…We also compare ESBMC-GPU to GKLEE [8], GPUVerify [3], PUG [7], and CIVL [9]. We also compare ESBMC-GPU to GKLEE [8], GPUVerify [3], PUG [7], and CIVL [9].…”

Section: Methodsmentioning

confidence: 99%

“…Thus, using operational models that simulate CUDA libraries, together with MPOR implementation in ESBMC-GPU, we achieve significant (correct) results of CUDA kernels verification, primarily when compared with other state-of-the-art GPU verifiers [3,[7][8][9]. ESBMC-GPU uses an operational model, that is, an abstract representation of the standard CUDA libraries, which conservatively approximates their semantics, in order to verify CUDA-based programs.…”

mentioning

confidence: 99%

See 2 more Smart Citations

SMT‐based context‐bounded model checking for CUDA programs

Pereira

Albuquerque

Silva

et al. 2016

Concurrency and Computation

View full text Add to dashboard Cite

We present ESBMC-GPU tool, an extension to the Efficient SMT-Based Context-Bounded Model Checker (ESBMC), which is aimed at verifying Graphics Processing Unit (GPU) programs written for the Compute Unified Device Architecture (CUDA) platform. ESBMC-GPU uses an operational model, that is, an abstract representation of the standard CUDA libraries, which conservatively approximates their semantics, in order to verify CUDA-based programs. It then explicitly explores the possible interleavings (up to the given context bound), while treats each interleaving itself symbolically. Additionally, ESBMC-GPU employs the monotonic partial order reduction and the two-thread analysis to prune the state space exploration. Experimental results show that ESBMC-GPU can successfully verify 82% of all benchmarks, while keeping lower rates of false results. Going further than previous attempts, ESBMC-GPU is able to detect more properties violations than other existing GPU verifiers due to its ability to verify errors of the program execution flow and to detect array out-of-bounds and data race violations. exploration, similar to Cordeiro et al. [4]. In particular, we explicitly explore the possible interleavings (up to the given context bound), while we treat each interleaving itself symbolically w.r.t. a given property.To prune the state-space exploration, we apply Monotonic Partial Order Reduction (MPOR) [12] to CUDA programs, which eliminates redundant interleavings without missing any behavior that can be exhibited by the program. We have modified the MPOR algorithm to identify transitions between threads that accessed different memory locations in the same array. Because CUDA kernels typically produce regular and independent access to explore the benefits of the GPU execution model, the application of MPOR routinely leads to substantial performance improvements in most benchmarks. Thus, using operational models that simulate CUDA libraries, together with MPOR implementation in ESBMC-GPU, we achieve significant (correct) results of CUDA kernels verification, primarily when compared with other state-of-the-art GPU verifiers [3,[7][8][9]. Additionally, the present approach considers low-level aspects related to dynamic memory allocation, data transfer, memory deallocation, and overflow. Such violations are typically present in CUDA programs, however, they are routinely ignored by most GPU verifiers. Thus, we provide a more precise verification than other existing approaches, considering soundness of data passed by the main program to the kernel, with the drawback of leading to a higher verification time. ContributionsWe make four major contributions:we extend benefits of SMT-based context-bounded model checking for CUDA programs, in the context of parallel programming for GPUs, to detect more failures than other existing approaches, while keeping lower rates of false results; although SMT-based context-bounded model checking is not a novel technique, we have not seen in the literature its application to verify CUDA programs. this wor...

show abstract

Section: Resultsmentioning

confidence: 99%

“…We also compare ESBMC-GPU to GKLEE [8], GPUVerify [3], PUG [7], and CIVL [9]. We also compare ESBMC-GPU to GKLEE [8], GPUVerify [3], PUG [7], and CIVL [9].…”

Section: Methodsmentioning

confidence: 99%

mentioning

confidence: 99%

See 1 more Smart Citation

SMT‐based context‐bounded model checking for CUDA programs

Pereira

Albuquerque

Silva

et al. 2016

Concurrency and Computation

View full text Add to dashboard Cite

show abstract

“…Recent advances for verifying multi-threaded C programs have been proposed to speed up the verification time, which significantly prune the state-space exploration [35,40]; however, the class of concurrent programs (e.g., CUDA, OpenCL, and MPI) that can be verified is still very limited. One possible research direction is to further extend BMC of multi-threaded C programs via Lazy Sequentialization [35], in order to analyze unsatisfiability cores [41] with the goal to remove redundant behaviour or to analyze interpolants [42] to prove noninterference of context switches.…”

Section: Current Achievements and Future Trendsmentioning

confidence: 99%

SMT-Based Context-Bounded Model Checking for Embedded Systems

Cordeiro

Filho

2016

SIGSOFT Softw. Eng. Notes

View full text Add to dashboard Cite

The dependency on the correct functioning of embedded systems is rapidly growing, mainly due to their wide range of applications, such as micro-grids, automotive device control (e.g., airbag control), health care, surveillance, mobile devices, and consumer electronics. Their structures are becoming more and more complex and now require multi-core processors with scalable shared memory, in order to meet increasing computational power demands. As a consequence, reliability of embedded (distributed) software becomes a key issue during system development, which must be carefully addressed and assured. Normally, state-of-the-art verification methodologies for embedded systems generate test vectors (with constraints) and use assertion-based verification and highlevel processor models, during simulation; however, other additional challenges have been raised: the need for meeting time and energy constraints, handling concurrent software, dealing with platform restrictions, evaluating implementation-structure choices, and supporting new software architectures and legacy designs (usually written in low-level languages). The present paper discusses challenges, problems, and recent advances to ensure correctness and timeliness regarding embedded systems. Reliability issues, in the development of micro-grids and cyber-physical systems, are then considered, as a prominent (bounded) model checking application.

show abstract

“…It contains a rare bug that is hard to find with automatic bug-finding techniques already under SC (including random testing, Nidhugg, CIVL [31], CBMC and other approaches based on BMC) [26]. The only tool we are aware of that can automatically find a genuine counter-example is Lazy-CSeq.…”

Section: Experimental Evaluationmentioning

confidence: 99%

Lazy sequentialization for TSO and PSO via shared memory abstractions

Tomasco¹,

Nguyen²,

Inverso³

et al. 2016

2016 Formal Methods in Computer-Aided Design (FMCAD)

View full text Add to dashboard Cite

Abstract-Lazy sequentialization is one of the most effective approaches for the bounded verification of concurrent programs. Existing tools assume sequential consistency (SC), thus the feasibility of lazy sequentializations for weak memory models (WMMs) remains untested. Here, we describe the first lazy sequentialization approach for the total store order (TSO) and partial store order (PSO) memory models. We replace all shared memory accesses with operations on a shared memory abstraction (SMA), an abstract data type that encapsulates the semantics of the underlying WMM and implements it under the simpler SC model. We give efficient SMA implementations for TSO and PSO that are based on temporal circular doubly-linked lists, a new data structure that allows an efficient simulation of the store buffers. We show experimentally, both on the SV-COMP concurrency benchmarks and a real world instance, that this approach works well in combination with lazy sequentialization on top of bounded model checking.

show abstract

CIVL: Formal Verification of Parallel Programs

Cited by 18 publications

References 15 publications

SMT‐based context‐bounded model checking for CUDA programs

SMT‐based context‐bounded model checking for CUDA programs

SMT-Based Context-Bounded Model Checking for Embedded Systems

Lazy sequentialization for TSO and PSO via shared memory abstractions

Contact Info

Product

Resources

About